[PATCH] D37540: [CUDA] Tests for device-side overloads of non-placement new/delete.

[Justin Lebar via Phabricator via llvm-commits]

jlebar created this revision.
Herald added subscribers: mgorny, sanjoy.

https://reviews.llvm.org/D37540

Files:
  External/CUDA/CMakeLists.txt
  External/CUDA/new.cu
  External/CUDA/new.reference_output


Index: External/CUDA/new.reference_output
===================================================================
--- /dev/null
+++ External/CUDA/new.reference_output
@@ -0,0 +1,2 @@
+Success!
+exit 0
Index: External/CUDA/new.cu
===================================================================
--- /dev/null
+++ External/CUDA/new.cu
@@ -0,0 +1,69 @@
+// Check that operator new and operator delete work.
+
+#include <assert.h>
+#include <new>
+#include <stdio.h>
+
+__device__ void global_new() {
+  void* x = ::operator new(42);
+  assert(x != nullptr);
+  ::operator delete(x);
+
+  x = ::operator new(42, std::nothrow);
+  assert(x != nullptr);
+  ::operator delete(x, std::nothrow);
+
+  x = ::operator new[](42);
+  assert(x != nullptr);
+  ::operator delete[](x);
+
+  x = ::operator new[](42, std::nothrow);
+  assert(x != nullptr);
+  ::operator delete[](x, std::nothrow);
+}
+
+__device__ void sized_delete() {
+#if __cplusplus>= 201402L
+  void* x = ::operator new(42);
+  assert(x != nullptr);
+  ::operator delete(x, 42);
+
+  x = ::operator new[](42);
+  assert(x != nullptr);
+  ::operator delete[](x, 42);
+#endif
+}
+
+__device__ void int_new() {
+  int* x = new int();
+  assert(*x == 0);
+  delete x;
+}
+
+struct Foo {
+  __device__ Foo() : x(42) {}
+  int x;
+};
+__device__ void class_new() {
+  Foo* foo = new Foo();
+  assert(foo->x == 42);
+  delete foo;
+}
+
+__global__ void kernel() {
+  global_new();
+  sized_delete();
+  int_new();
+  class_new();
+}
+
+int main() {
+  kernel<<<32, 32>>>();
+  cudaError_t err = cudaDeviceSynchronize();
+  if (err != cudaSuccess) {
+    printf("CUDA error %d\n", (int)err);
+    return 1;
+  }
+  printf("Success!\n");
+  return 0;
+}
Index: External/CUDA/CMakeLists.txt
===================================================================
--- External/CUDA/CMakeLists.txt
+++ External/CUDA/CMakeLists.txt
@@ -53,6 +53,7 @@
   create_one_local_test(cmath cmath.cu)
   create_one_local_test(complex complex.cu)
   create_one_local_test(math_h math_h.cu)
+  create_one_local_test(new new.cu)
   create_one_local_test(empty empty.cu)
   create_one_local_test(printf printf.cu)
   create_one_local_test(future future.cu)


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D37540.114094.patch
Type: text/x-patch
Size: 2180 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/llvm-commits/attachments/20170906/e58a23d5/attachment.bin>