44 lines · c
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-exception6 *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#endif14 15#if !defined(__cplusplus)16 17#include <limits.h>18 19#ifdef __OPENMP_AMDGCN__20#define __DEVICE__ static inline __attribute__((always_inline, nothrow))21#else22#define __DEVICE__ static __device__ inline __attribute__((always_inline))23#endif24 25__DEVICE__26int abs(int __x) {27 int __sgn = __x >> (sizeof(int) * CHAR_BIT - 1);28 return (__x ^ __sgn) - __sgn;29}30__DEVICE__31long labs(long __x) {32 long __sgn = __x >> (sizeof(long) * CHAR_BIT - 1);33 return (__x ^ __sgn) - __sgn;34}35__DEVICE__36long 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