1 /*===- __clang_openmp_device_functions.h - OpenMP device function declares -=== 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 #ifndef __CLANG_OPENMP_DEVICE_FUNCTIONS_H__ 11 #define __CLANG_OPENMP_DEVICE_FUNCTIONS_H__ 12 13 #ifndef _OPENMP 14 #error "This file is for OpenMP compilation only." 15 #endif 16 17 #ifdef __cplusplus 18 extern "C" { 19 #endif 20 21 #pragma omp begin declare variant match( \ 22 device = {arch(nvptx, nvptx64)}, implementation = {extension(match_any)}) 23 24 #define __CUDA__ 25 #define __OPENMP_NVPTX__ 26 27 /// Include declarations for libdevice functions. 28 #include <__clang_cuda_libdevice_declares.h> 29 30 /// Provide definitions for these functions. 31 #include <__clang_cuda_device_functions.h> 32 33 #undef __OPENMP_NVPTX__ 34 #undef __CUDA__ 35 36 #pragma omp end declare variant 37 38 #pragma omp begin declare variant match(device = {arch(amdgcn)}) 39 40 // Import types which will be used by __clang_hip_libdevice_declares.h 41 #ifndef __cplusplus 42 #include <stdbool.h> 43 #include <stdint.h> 44 #endif 45 46 #define __OPENMP_AMDGCN__ 47 #pragma push_macro("__device__") 48 #define __device__ 49 50 /// Include declarations for libdevice functions. 51 #include <__clang_hip_libdevice_declares.h> 52 53 #pragma pop_macro("__device__") 54 #undef __OPENMP_AMDGCN__ 55 56 #pragma omp end declare variant 57 58 #ifdef __cplusplus 59 } // extern "C" 60 #endif 61 62 // Ensure we make `_ZdlPv`, aka. `operator delete(void*)` available without the 63 // need to `include <new>` in C++ mode. 64 #ifdef __cplusplus 65 66 // We require malloc/free. 67 #include <cstdlib> 68 69 #pragma push_macro("OPENMP_NOEXCEPT") 70 #if __cplusplus >= 201103L 71 #define OPENMP_NOEXCEPT noexcept 72 #else 73 #define OPENMP_NOEXCEPT 74 #endif 75 76 // Device overrides for non-placement new and delete. 77 inline void *operator new(__SIZE_TYPE__ size) { 78 if (size == 0) 79 size = 1; 80 return ::malloc(size); 81 } 82 83 inline void *operator new[](__SIZE_TYPE__ size) { return ::operator new(size); } 84 85 inline void operator delete(void *ptr)OPENMP_NOEXCEPT { ::free(ptr); } 86 87 inline void operator delete[](void *ptr) OPENMP_NOEXCEPT { 88 ::operator delete(ptr); 89 } 90 91 // Sized delete, C++14 only. 92 #if __cplusplus >= 201402L 93 inline void operator delete(void *ptr, __SIZE_TYPE__ size)OPENMP_NOEXCEPT { 94 ::operator delete(ptr); 95 } 96 inline void operator delete[](void *ptr, __SIZE_TYPE__ size) OPENMP_NOEXCEPT { 97 ::operator delete(ptr); 98 } 99 #endif 100 101 #pragma pop_macro("OPENMP_NOEXCEPT") 102 #endif 103 104 #endif 105