Author: jlebar Date: Wed Feb 24 15:49:31 2016 New Revision: 261776 URL: http://llvm.org/viewvc/llvm-project?rev=261776&view=rev Log: [CUDA] Add hack so code which includes "curand.h" doesn't break.
Summary: curand.h includes curand_mtgp32_kernel.h. In host mode, this header redefines threadIdx and blockDim, giving them their "proper" types of uint3 and dim3, respectively. clang has its own plan for these variables -- their types are magic builtin classes. So these redefinitions are incompatible. As a hack, we force-include the offending CUDA header and use #defines to get the right types for threadIdx and blockDim. Reviewers: tra Subscribers: echristo, cfe-commits Differential Revision: http://reviews.llvm.org/D17562 Modified: cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h Modified: cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h URL: http://llvm.org/viewvc/llvm-project/cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h?rev=261776&r1=261775&r2=261776&view=diff ============================================================================== --- cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h (original) +++ cfe/trunk/lib/Headers/__clang_cuda_runtime_wrapper.h Wed Feb 24 15:49:31 2016 @@ -247,5 +247,19 @@ __device__ static inline void *malloc(si #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__ _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org http://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
