Subversion Repositories QNX 8.QNX8 LLVM/Clang compiler suite

Rev

Blame | Last modification | View Log | Download | RSS feed

  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. #include <__clang_hip_stdlib.h>
  117.  
  118. #if defined(__HIPCC_RTC__)
  119. #include <__clang_hip_cmath.h>
  120. #else
  121. #include <__clang_cuda_math_forward_declares.h>
  122. #include <__clang_hip_cmath.h>
  123. #include <__clang_cuda_complex_builtins.h>
  124. #include <algorithm>
  125. #include <complex>
  126. #include <new>
  127. #endif // __HIPCC_RTC__
  128.  
  129. #define __CLANG_HIP_RUNTIME_WRAPPER_INCLUDED__ 1
  130. #if defined(__HIPCC_RTC__)
  131. #pragma pop_macro("NULL")
  132. #pragma pop_macro("uint32_t")
  133. #pragma pop_macro("uint64_t")
  134. #pragma pop_macro("CHAR_BIT")
  135. #pragma pop_macro("INT_MAX")
  136. #endif // __HIPCC_RTC__
  137. #endif // __HIP__
  138. #endif // __CLANG_HIP_RUNTIME_WRAPPER_H__
  139.