[PATCH] D17562: [CUDA] Add hack so code which includes "curand.h" doesn't break.

[Justin Lebar via cfe-commits]

This revision was automatically updated to reflect the committed changes.
Closed by commit rL261776: [CUDA] Add hack so code which includes "curand.h" doesn't break. (authored by jlebar).

Changed prior to commit:
  http://reviews.llvm.org/D17562?vs=48885&id=48977#toc

Repository:
  rL LLVM

http://reviews.llvm.org/D17562

Files:
  cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h

Index: cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h
===================================================================
--- cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h
+++ cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h
@@ -247,5 +247,19 @@
 
 #include <__clang_cuda_cmath.h>
 
+// curand_mtgp32_kernel helpfully redeclares blockDim and threadIdx in host
+// mode, giving them their "proper" types of dim3 and uint3.  This is
+// incompatible with the types we give in cuda_builtin_vars.h.  As as hack,
+// force-include the header (nvcc doesn't include it by default) but redefine
+// dim3 and uint3 to our builtin types.  (Thankfully dim3 and uint3 are only
+// used here for the redeclarations of blockDim and threadIdx.)
+#pragma push_macro("dim3")
+#pragma push_macro("uint3")
+#define dim3 __cuda_builtin_blockDim_t
+#define uint3 __cuda_builtin_threadIdx_t
+#include "curand_mtgp32_kernel.h"
+#pragma pop_macro("dim3")
+#pragma pop_macro("uint3")
+
 #endif // __CUDA__
 #endif // __CLANG_CUDA_RUNTIME_WRAPPER_H__


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D17562.48977.patch
Type: text/x-patch
Size: 1042 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20160224/24e14cc5/attachment.bin>