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#ifdef __cplusplus
 14extern "C" {
 15#endif
 16
 17#ifdef __NVPTX__
 18#pragma omp begin declare variant match(                                       \
 19    device = {arch(nvptx, nvptx64)}, implementation = {extension(match_any)})
 20
 21#pragma push_macro("__CUDA__")
 22#define __CUDA__
 23#define __OPENMP_NVPTX__
 24
 25/// Include declarations for libdevice functions.
 26#include <__clang_cuda_libdevice_declares.h>
 27
 28/// Provide definitions for these functions.
 29#include <__clang_cuda_device_functions.h>
 30
 31#undef __OPENMP_NVPTX__
 32#pragma pop_macro("__CUDA__")
 33
 34#pragma omp end declare variant
 35#endif
 36
 37#ifdef __AMDGCN__
 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 <stdint.h>
 43#endif
 44
 45#define __OPENMP_AMDGCN__
 46#pragma push_macro("__device__")
 47#define __device__
 48
 49/// Include declarations for libdevice functions.
 50#include <__clang_hip_libdevice_declares.h>
 51
 52#pragma pop_macro("__device__")
 53#undef __OPENMP_AMDGCN__
 54
 55#pragma omp end declare variant
 56#endif
 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.
 77inline void *operator new(__SIZE_TYPE__ size) {
 78  if (size == 0)
 79    size = 1;
 80  return ::malloc(size);
 81}
 82
 83inline void *operator new[](__SIZE_TYPE__ size) { return ::operator new(size); }
 84
 85inline void operator delete(void *ptr)OPENMP_NOEXCEPT { ::free(ptr); }
 86
 87inline void operator delete[](void *ptr) OPENMP_NOEXCEPT {
 88  ::operator delete(ptr);
 89}
 90
 91// Sized delete, C++14 only.
 92#if __cplusplus >= 201402L
 93inline void operator delete(void *ptr, __SIZE_TYPE__ size)OPENMP_NOEXCEPT {
 94  ::operator delete(ptr);
 95}
 96inline 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