[PATCH] D16638: [CUDA] Added device-side std::{malloc/free}

Artem Belevich via cfe-commits cfe-commits at lists.llvm.org
Wed Jan 27 10:28:04 PST 2016

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

In addition to math functions, we also need to support std::malloc and std::free to match NVCC behavior.



Index: lib/Headers/__clang_cuda_runtime_wrapper.h
--- lib/Headers/__clang_cuda_runtime_wrapper.h
+++ lib/Headers/__clang_cuda_runtime_wrapper.h
@@ -169,6 +169,10 @@
 // CUDA headers. Alas, device_functions.hpp included below needs it.
 static inline __device__ void __brkpt(int c) { __brkpt(); }
+// We also need extern "C" decls for device-side allocator functions.
+extern "C" __device__ void free(void *__ptr);
+extern "C" __device__ void *malloc(size_t __size);
 // Now include *.hpp with definitions of various GPU functions.  Alas,
 // a lot of thins get declared/defined with __host__ attribute which
 // we don't want and we have to define it out. We also have to include
Index: lib/Headers/__clang_cuda_cmath.h
--- lib/Headers/__clang_cuda_cmath.h
+++ lib/Headers/__clang_cuda_cmath.h
@@ -218,6 +218,9 @@
 __DEVICE__ float trunc(float x) { return ::truncf(x); }
 __DEVICE__ double trunc(double x) { return ::trunc(x); }
+__DEVICE__ void free(void *__ptr) { return ::free(__ptr); }
+__DEVICE__ void *malloc(size_t __size) { return ::malloc(__size); }
 } // namespace std

-------------- next part --------------
A non-text attachment was scrubbed...
Name: D16638.46151.patch
Type: text/x-patch
Size: 1233 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20160127/3c14820e/attachment.bin>

More information about the cfe-commits mailing list