1 /*===---- __clang_hip_runtime_wrapper.h - HIP runtime 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 10 /* 11 * WARNING: This header is intended to be directly -include'd by 12 * the compiler and is not supposed to be included by users. 13 * 14 */ 15 16 #ifndef __CLANG_HIP_RUNTIME_WRAPPER_H__ 17 #define __CLANG_HIP_RUNTIME_WRAPPER_H__ 18 19 #if __HIP__ 20 21 #define __host__ __attribute__((host)) 22 #define __device__ __attribute__((device)) 23 #define __global__ __attribute__((global)) 24 #define __shared__ __attribute__((shared)) 25 #define __constant__ __attribute__((constant)) 26 #define __managed__ __attribute__((managed)) 27 28 #if !defined(__cplusplus) || __cplusplus < 201103L 29 #define nullptr NULL; 30 #endif 31 32 #ifdef __cplusplus 33 extern "C" { 34 __attribute__((__visibility__("default"))) 35 __attribute__((weak)) 36 __attribute__((noreturn)) 37 __device__ void __cxa_pure_virtual(void) { 38 __builtin_trap(); 39 } 40 __attribute__((__visibility__("default"))) 41 __attribute__((weak)) 42 __attribute__((noreturn)) 43 __device__ void __cxa_deleted_virtual(void) { 44 __builtin_trap(); 45 } 46 } 47 #endif //__cplusplus 48 49 #if !defined(__HIPCC_RTC__) 50 #include <cmath> 51 #include <cstdlib> 52 #include <stdlib.h> 53 #if __has_include("hip/hip_version.h") 54 #include "hip/hip_version.h" 55 #endif // __has_include("hip/hip_version.h") 56 #else 57 typedef __SIZE_TYPE__ size_t; 58 // Define macros which are needed to declare HIP device API's without standard 59 // C/C++ headers. This is for readability so that these API's can be written 60 // the same way as non-hipRTC use case. These macros need to be popped so that 61 // they do not pollute users' name space. 62 #pragma push_macro("NULL") 63 #pragma push_macro("uint32_t") 64 #pragma push_macro("uint64_t") 65 #pragma push_macro("CHAR_BIT") 66 #pragma push_macro("INT_MAX") 67 #define NULL (void *)0 68 #define uint32_t __UINT32_TYPE__ 69 #define uint64_t __UINT64_TYPE__ 70 #define CHAR_BIT __CHAR_BIT__ 71 #define INT_MAX __INTMAX_MAX__ 72 #endif // __HIPCC_RTC__ 73 74 typedef __SIZE_TYPE__ __hip_size_t; 75 76 #ifdef __cplusplus 77 extern "C" { 78 #endif //__cplusplus 79 80 #if HIP_VERSION_MAJOR * 100 + HIP_VERSION_MINOR >= 405 81 extern "C" __device__ unsigned long long __ockl_dm_alloc(unsigned long long __size); 82 extern "C" __device__ void __ockl_dm_dealloc(unsigned long long __addr); 83 __attribute__((weak)) inline __device__ void *malloc(__hip_size_t __size) { 84 return (void *) __ockl_dm_alloc(__size); 85 } 86 __attribute__((weak)) inline __device__ void free(void *__ptr) { 87 __ockl_dm_dealloc((unsigned long long)__ptr); 88 } 89 #else // HIP version check 90 #if __HIP_ENABLE_DEVICE_MALLOC__ 91 __device__ void *__hip_malloc(__hip_size_t __size); 92 __device__ void *__hip_free(void *__ptr); 93 __attribute__((weak)) inline __device__ void *malloc(__hip_size_t __size) { 94 return __hip_malloc(__size); 95 } 96 __attribute__((weak)) inline __device__ void free(void *__ptr) { 97 __hip_free(__ptr); 98 } 99 #else 100 __attribute__((weak)) inline __device__ void *malloc(__hip_size_t __size) { 101 __builtin_trap(); 102 return (void *)0; 103 } 104 __attribute__((weak)) inline __device__ void free(void *__ptr) { 105 __builtin_trap(); 106 } 107 #endif 108 #endif // HIP version check 109 110 #ifdef __cplusplus 111 } // extern "C" 112 #endif //__cplusplus 113 114 #include <__clang_hip_libdevice_declares.h> 115 #include <__clang_hip_math.h> 116 117 #if defined(__HIPCC_RTC__) 118 #include <__clang_hip_cmath.h> 119 #else 120 #include <__clang_cuda_math_forward_declares.h> 121 #include <__clang_hip_cmath.h> 122 #include <__clang_cuda_complex_builtins.h> 123 #include <algorithm> 124 #include <complex> 125 #include <new> 126 #endif // __HIPCC_RTC__ 127 128 #define __CLANG_HIP_RUNTIME_WRAPPER_INCLUDED__ 1 129 #if defined(__HIPCC_RTC__) 130 #pragma pop_macro("NULL") 131 #pragma pop_macro("uint32_t") 132 #pragma pop_macro("uint64_t") 133 #pragma pop_macro("CHAR_BIT") 134 #pragma pop_macro("INT_MAX") 135 #endif // __HIPCC_RTC__ 136 #endif // __HIP__ 137 #endif // __CLANG_HIP_RUNTIME_WRAPPER_H__ 138