1 /*===---- __clang_hip_stdlib.h - Device-side HIP math support --------------===
2  *
3  * Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4  * See https://llvm.org/LICENSE.txt for license information.
5  * SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6  *
7  *===-----------------------------------------------------------------------===
8  */
9 #ifndef __CLANG_HIP_STDLIB_H__
10 
11 #if !defined(__HIP__) && !defined(__OPENMP_AMDGCN__)
12 #error "This file is for HIP and OpenMP AMDGCN device compilation only."
13 #endif
14 
15 #if !defined(__cplusplus)
16 
17 #include <limits.h>
18 
19 #ifdef __OPENMP_AMDGCN__
20 #define __DEVICE__ static inline __attribute__((always_inline, nothrow))
21 #else
22 #define __DEVICE__ static __device__ inline __attribute__((always_inline))
23 #endif
24 
25 __DEVICE__
abs(int __x)26 int abs(int __x) {
27   int __sgn = __x >> (sizeof(int) * CHAR_BIT - 1);
28   return (__x ^ __sgn) - __sgn;
29 }
30 __DEVICE__
labs(long __x)31 long labs(long __x) {
32   long __sgn = __x >> (sizeof(long) * CHAR_BIT - 1);
33   return (__x ^ __sgn) - __sgn;
34 }
35 __DEVICE__
llabs(long long __x)36 long long llabs(long long __x) {
37   long long __sgn = __x >> (sizeof(long long) * CHAR_BIT - 1);
38   return (__x ^ __sgn) - __sgn;
39 }
40 
41 #endif // !defined(__cplusplus)
42 
43 #endif // #define __CLANG_HIP_STDLIB_H__
44