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__ 26 int abs(int __x) { 27 int __sgn = __x >> (sizeof(int) * CHAR_BIT - 1); 28 return (__x ^ __sgn) - __sgn; 29 } 30 __DEVICE__ 31 long labs(long __x) { 32 long __sgn = __x >> (sizeof(long) * CHAR_BIT - 1); 33 return (__x ^ __sgn) - __sgn; 34 } 35 __DEVICE__ 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