Home | History | Annotate | Line # | Download | only in cuda_wrappers 
      1 /*===---- new - CUDA wrapper for <new> -------------------------------------===
      2  *
      3  * Permission is hereby granted, free of charge, to any person obtaining a copy
      4  * of this software and associated documentation files (the "Software"), to deal
      5  * in the Software without restriction, including without limitation the rights
      6  * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
      7  * copies of the Software, and to permit persons to whom the Software is
      8  * furnished to do so, subject to the following conditions:
      9  *
     10  * The above copyright notice and this permission notice shall be included in
     11  * all copies or substantial portions of the Software.
     12  *
     13  * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
     14  * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
     15  * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
     16  * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
     17  * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
     18  * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
     19  * THE SOFTWARE.
     20  *
     21  *===-----------------------------------------------------------------------===
     22  */
     23 
     24 #ifndef __CLANG_CUDA_WRAPPERS_NEW
     25 #define __CLANG_CUDA_WRAPPERS_NEW
     26 
     27 #include_next <new>
     28 
     29 #if !defined(__device__)
     30 // The header has been included too early from the standard C++ library
     31 // and CUDA-specific macros are not available yet.
     32 // Undo the include guard and try again later.
     33 #undef __CLANG_CUDA_WRAPPERS_NEW
     34 #else
     35 
     36 #pragma push_macro("CUDA_NOEXCEPT")
     37 #if __cplusplus >= 201103L
     38 #define CUDA_NOEXCEPT noexcept
     39 #else
     40 #define CUDA_NOEXCEPT
     41 #endif
     42 
     43 // Device overrides for non-placement new and delete.
     44 __device__ inline void *operator new(__SIZE_TYPE__ size) {
     45   if (size == 0) {
     46     size = 1;
     47   }
     48   return ::malloc(size);
     49 }
     50 __device__ inline void *operator new(__SIZE_TYPE__ size,
     51                                      const std::nothrow_t &) CUDA_NOEXCEPT {
     52   return ::operator new(size);
     53 }
     54 
     55 __device__ inline void *operator new[](__SIZE_TYPE__ size) {
     56   return ::operator new(size);
     57 }
     58 __device__ inline void *operator new[](__SIZE_TYPE__ size,
     59                                        const std::nothrow_t &) {
     60   return ::operator new(size);
     61 }
     62 
     63 __device__ inline void operator delete(void* ptr) CUDA_NOEXCEPT {
     64   if (ptr) {
     65     ::free(ptr);
     66   }
     67 }
     68 __device__ inline void operator delete(void *ptr,
     69                                        const std::nothrow_t &) CUDA_NOEXCEPT {
     70   ::operator delete(ptr);
     71 }
     72 
     73 __device__ inline void operator delete[](void* ptr) CUDA_NOEXCEPT {
     74   ::operator delete(ptr);
     75 }
     76 __device__ inline void operator delete[](void *ptr,
     77                                          const std::nothrow_t &) CUDA_NOEXCEPT {
     78   ::operator delete(ptr);
     79 }
     80 
     81 // Sized delete, C++14 only.
     82 #if __cplusplus >= 201402L
     83 __device__ inline void operator delete(void *ptr,
     84                                        __SIZE_TYPE__ size) CUDA_NOEXCEPT {
     85   ::operator delete(ptr);
     86 }
     87 __device__ inline void operator delete[](void *ptr,
     88                                          __SIZE_TYPE__ size) CUDA_NOEXCEPT {
     89   ::operator delete(ptr);
     90 }
     91 #endif
     92 
     93 // Device overrides for placement new and delete.
     94 __device__ inline void *operator new(__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT {
     95   return __ptr;
     96 }
     97 __device__ inline void *operator new[](__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT {
     98   return __ptr;
     99 }
    100 __device__ inline void operator delete(void *, void *) CUDA_NOEXCEPT {}
    101 __device__ inline void operator delete[](void *, void *) CUDA_NOEXCEPT {}
    102 
    103 #pragma pop_macro("CUDA_NOEXCEPT")
    104 
    105 #endif // __device__
    106 #endif // include guard
    107