This revision was automatically updated to reflect the committed changes.
Closed by commit rL312681: [CUDA] Add device overloads for non-placement
new/delete. (authored by jlebar).
Changed prior to commit:
https://reviews.llvm.org/D37539?vs=114104&id=114108#toc
Repository:
rL LLVM
https://reviews.llvm.org/D37539
Files:
cfe/trunk/lib/Headers/cuda_wrappers/new
Index: cfe/trunk/lib/Headers/cuda_wrappers/new
===================================================================
--- cfe/trunk/lib/Headers/cuda_wrappers/new
+++ cfe/trunk/lib/Headers/cuda_wrappers/new
@@ -26,22 +26,71 @@
#include_next <new>
-// Device overrides for placement new and delete.
#pragma push_macro("CUDA_NOEXCEPT")
#if __cplusplus >= 201103L
#define CUDA_NOEXCEPT noexcept
#else
#define CUDA_NOEXCEPT
#endif
+// Device overrides for non-placement new and delete.
+__device__ inline void *operator new(__SIZE_TYPE__ size) {
+ if (size == 0) {
+ size = 1;
+ }
+ return ::malloc(size);
+}
+__device__ inline void *operator new(__SIZE_TYPE__ size,
+ const std::nothrow_t &) CUDA_NOEXCEPT {
+ return ::operator new(size);
+}
+
+__device__ inline void *operator new[](__SIZE_TYPE__ size) {
+ return ::operator new(size);
+}
+__device__ inline void *operator new[](__SIZE_TYPE__ size,
+ const std::nothrow_t &) {
+ return ::operator new(size);
+}
+
+__device__ inline void operator delete(void* ptr) CUDA_NOEXCEPT {
+ if (ptr) {
+ ::free(ptr);
+ }
+}
+__device__ inline void operator delete(void *ptr,
+ const std::nothrow_t &) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+
+__device__ inline void operator delete[](void* ptr) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+__device__ inline void operator delete[](void *ptr,
+ const std::nothrow_t &) CUDA_NOEXCEPT
{
+ ::operator delete(ptr);
+}
+
+// Sized delete, C++14 only.
+#if __cplusplus >= 201402L
+__device__ void operator delete(void *ptr, __SIZE_TYPE__ size) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+__device__ void operator delete[](void *ptr, __SIZE_TYPE__ size) CUDA_NOEXCEPT
{
+ ::operator delete(ptr);
+}
+#endif
+
+// Device overrides for placement new and delete.
__device__ inline void *operator new(__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT
{
return __ptr;
}
__device__ inline void *operator new[](__SIZE_TYPE__, void *__ptr)
CUDA_NOEXCEPT {
return __ptr;
}
__device__ inline void operator delete(void *, void *) CUDA_NOEXCEPT {}
__device__ inline void operator delete[](void *, void *) CUDA_NOEXCEPT {}
+
#pragma pop_macro("CUDA_NOEXCEPT")
#endif // include guard
Index: cfe/trunk/lib/Headers/cuda_wrappers/new
===================================================================
--- cfe/trunk/lib/Headers/cuda_wrappers/new
+++ cfe/trunk/lib/Headers/cuda_wrappers/new
@@ -26,22 +26,71 @@
#include_next <new>
-// Device overrides for placement new and delete.
#pragma push_macro("CUDA_NOEXCEPT")
#if __cplusplus >= 201103L
#define CUDA_NOEXCEPT noexcept
#else
#define CUDA_NOEXCEPT
#endif
+// Device overrides for non-placement new and delete.
+__device__ inline void *operator new(__SIZE_TYPE__ size) {
+ if (size == 0) {
+ size = 1;
+ }
+ return ::malloc(size);
+}
+__device__ inline void *operator new(__SIZE_TYPE__ size,
+ const std::nothrow_t &) CUDA_NOEXCEPT {
+ return ::operator new(size);
+}
+
+__device__ inline void *operator new[](__SIZE_TYPE__ size) {
+ return ::operator new(size);
+}
+__device__ inline void *operator new[](__SIZE_TYPE__ size,
+ const std::nothrow_t &) {
+ return ::operator new(size);
+}
+
+__device__ inline void operator delete(void* ptr) CUDA_NOEXCEPT {
+ if (ptr) {
+ ::free(ptr);
+ }
+}
+__device__ inline void operator delete(void *ptr,
+ const std::nothrow_t &) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+
+__device__ inline void operator delete[](void* ptr) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+__device__ inline void operator delete[](void *ptr,
+ const std::nothrow_t &) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+
+// Sized delete, C++14 only.
+#if __cplusplus >= 201402L
+__device__ void operator delete(void *ptr, __SIZE_TYPE__ size) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+__device__ void operator delete[](void *ptr, __SIZE_TYPE__ size) CUDA_NOEXCEPT {
+ ::operator delete(ptr);
+}
+#endif
+
+// Device overrides for placement new and delete.
__device__ inline void *operator new(__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT {
return __ptr;
}
__device__ inline void *operator new[](__SIZE_TYPE__, void *__ptr) CUDA_NOEXCEPT {
return __ptr;
}
__device__ inline void operator delete(void *, void *) CUDA_NOEXCEPT {}
__device__ inline void operator delete[](void *, void *) CUDA_NOEXCEPT {}
+
#pragma pop_macro("CUDA_NOEXCEPT")
#endif // include guard
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
