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 18extern "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#ifdef __AMDGCN__ 39#pragma omp begin declare variant match(device = {arch(amdgcn)}) 40 41// Import types which will be used by __clang_hip_libdevice_declares.h 42#ifndef __cplusplus 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#endif 58 59#ifdef __cplusplus 60} // extern "C" 61#endif 62 63// Ensure we make `_ZdlPv`, aka. `operator delete(void*)` available without the 64// need to `include <new>` in C++ mode. 65#ifdef __cplusplus 66 67// We require malloc/free. 68#include <cstdlib> 69 70#pragma push_macro("OPENMP_NOEXCEPT") 71#if __cplusplus >= 201103L 72#define OPENMP_NOEXCEPT noexcept 73#else 74#define OPENMP_NOEXCEPT 75#endif 76 77// Device overrides for non-placement new and delete. 78inline void *operator new(__SIZE_TYPE__ size) { 79 if (size == 0) 80 size = 1; 81 return ::malloc(size); 82} 83 84inline void *operator new[](__SIZE_TYPE__ size) { return ::operator new(size); } 85 86inline void operator delete(void *ptr)OPENMP_NOEXCEPT { ::free(ptr); } 87 88inline void operator delete[](void *ptr) OPENMP_NOEXCEPT { 89 ::operator delete(ptr); 90} 91 92// Sized delete, C++14 only. 93#if __cplusplus >= 201402L 94inline void operator delete(void *ptr, __SIZE_TYPE__ size)OPENMP_NOEXCEPT { 95 ::operator delete(ptr); 96} 97inline void operator delete[](void *ptr, __SIZE_TYPE__ size) OPENMP_NOEXCEPT { 98 ::operator delete(ptr); 99} 100#endif 101 102#pragma pop_macro("OPENMP_NOEXCEPT") 103#endif 104 105#endif 106