[PATCH] D23239: [CUDA] Add __device__ overloads for placement new and delete.

Justin Lebar via cfe-commits cfe-commits at lists.llvm.org
Sat Aug 6 12:36:41 PDT 2016

jlebar created this revision.
jlebar added a reviewer: tra.
jlebar added a subscriber: cfe-commits.

Previously these sort of worked because they didn't end up resulting in
calls at the ptx layer.  But I'm adding stricter checks that break
placement new without these changes.



Index: clang/lib/Headers/__clang_cuda_runtime_wrapper.h
--- clang/lib/Headers/__clang_cuda_runtime_wrapper.h
+++ clang/lib/Headers/__clang_cuda_runtime_wrapper.h
@@ -312,5 +312,15 @@
 #pragma pop_macro("uint3")
 #pragma pop_macro("__USE_FAST_MATH__")
+// Device overrides for placement new and delete.
+__device__ inline void *operator new(__SIZE_TYPE__, void *__ptr) {
+  return __ptr;
+__device__ inline void *operator new[](__SIZE_TYPE__, void *__ptr) {
+  return __ptr;
+__device__ inline void operator delete(void *, void *) {}
+__device__ inline void operator delete[](void *, void *) {}
 #endif // __CUDA__

-------------- next part --------------
A non-text attachment was scrubbed...
Name: D23239.67080.patch
Type: text/x-patch
Size: 736 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20160806/dc2f6480/attachment.bin>

More information about the cfe-commits mailing list