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))
__cxa_pure_virtual(void)37   __device__ void __cxa_pure_virtual(void) {
38     __builtin_trap();
39   }
40   __attribute__((__visibility__("default")))
41   __attribute__((weak))
42   __attribute__((noreturn))
__cxa_deleted_virtual(void)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 #else
54 typedef __SIZE_TYPE__ size_t;
55 // Define macros which are needed to declare HIP device API's without standard
56 // C/C++ headers. This is for readability so that these API's can be written
57 // the same way as non-hipRTC use case. These macros need to be popped so that
58 // they do not pollute users' name space.
59 #pragma push_macro("NULL")
60 #pragma push_macro("uint32_t")
61 #pragma push_macro("uint64_t")
62 #pragma push_macro("CHAR_BIT")
63 #pragma push_macro("INT_MAX")
64 #define NULL (void *)0
65 #define uint32_t __UINT32_TYPE__
66 #define uint64_t __UINT64_TYPE__
67 #define CHAR_BIT __CHAR_BIT__
68 #define INT_MAX __INTMAX_MAX__
69 #endif // __HIPCC_RTC__
70 
71 typedef __SIZE_TYPE__ __hip_size_t;
72 
73 #ifdef __cplusplus
74 extern "C" {
75 #endif //__cplusplus
76 
77 #if __HIP_ENABLE_DEVICE_MALLOC__
78 __device__ void *__hip_malloc(__hip_size_t __size);
79 __device__ void *__hip_free(void *__ptr);
malloc(__hip_size_t __size)80 __attribute__((weak)) inline __device__ void *malloc(__hip_size_t __size) {
81   return __hip_malloc(__size);
82 }
free(void * __ptr)83 __attribute__((weak)) inline __device__ void *free(void *__ptr) {
84   return __hip_free(__ptr);
85 }
86 #else
87 __attribute__((weak)) inline __device__ void *malloc(__hip_size_t __size) {
88   __builtin_trap();
89   return (void *)0;
90 }
91 __attribute__((weak)) inline __device__ void *free(void *__ptr) {
92   __builtin_trap();
93   return (void *)0;
94 }
95 #endif
96 
97 #ifdef __cplusplus
98 } // extern "C"
99 #endif //__cplusplus
100 
101 #include <__clang_hip_libdevice_declares.h>
102 #include <__clang_hip_math.h>
103 
104 #if defined(__HIPCC_RTC__)
105 #include <__clang_hip_cmath.h>
106 #else
107 #include <__clang_cuda_math_forward_declares.h>
108 #include <__clang_hip_cmath.h>
109 #include <__clang_cuda_complex_builtins.h>
110 #include <algorithm>
111 #include <complex>
112 #include <new>
113 #endif // __HIPCC_RTC__
114 
115 #define __CLANG_HIP_RUNTIME_WRAPPER_INCLUDED__ 1
116 #if defined(__HIPCC_RTC__)
117 #pragma pop_macro("NULL")
118 #pragma pop_macro("uint32_t")
119 #pragma pop_macro("uint64_t")
120 #pragma pop_macro("CHAR_BIT")
121 #pragma pop_macro("INT_MAX")
122 #endif // __HIPCC_RTC__
123 #endif // __HIP__
124 #endif // __CLANG_HIP_RUNTIME_WRAPPER_H__
125