diff options
| author | Evgeny Mankov <evgeny.mankov@gmail.com> | 2022-04-21 00:41:20 +0300 |
|---|---|---|
| committer | Tom Stellard <tstellar@redhat.com> | 2022-06-02 22:27:35 -0700 |
| commit | 79147e4722cc55b819b00007d0ac9a7d2f036956 (patch) | |
| tree | 9a02cf7b522ed933daecd2bfc899040893f49006 | |
| parent | ec0332328bd6311b740ce39e684ef3f9674edec2 (diff) | |
| download | llvm-79147e4722cc55b819b00007d0ac9a7d2f036956.tar.gz | |
[clang][CUDA][Windows] Fix compilation error on Windows with `uint32_t __nvvm_get_smem_pointer`
The change fixes https://github.com/llvm/llvm-project/issues/54609 (the second reported issue) by eliminating a compilation error occurring only on Windows while trying to compile any CUDA source file by clang (-x cuda).
[Repro]
clang -x cuda <any_cu_source>
[Error]
__clang_cuda_runtime_wrapper.h:473:
__clang_cuda_intrinsics.h(517,19): error GC871EEFB: unknown type name 'uint32_t'; did you mean 'cuuint32_t'?
__device__ inline uint32_t __nvvm_get_smem_pointer(void *__ptr) {
^
C:/Program Files/NVIDIA GPU Computing Toolkit/CUDA/v11.6/include\cuda.h:57:26: note: 'cuuint32_t' declared here
typedef unsigned __int32 cuuint32_t;
Reviewed By: tra
Differential Revision: https://reviews.llvm.org/D122897
(cherry picked from commit c23147106f7efc4b5e29c47a08951116b4d994ac)
| -rw-r--r-- | clang/lib/Headers/__clang_cuda_intrinsics.h | 2 |
1 files changed, 1 insertions, 1 deletions
diff --git a/clang/lib/Headers/__clang_cuda_intrinsics.h b/clang/lib/Headers/__clang_cuda_intrinsics.h index e0875bbcbf4a..3d2f5f40a435 100644 --- a/clang/lib/Headers/__clang_cuda_intrinsics.h +++ b/clang/lib/Headers/__clang_cuda_intrinsics.h @@ -509,7 +509,7 @@ __device__ inline void *__nv_cvta_constant_to_generic_impl(size_t __ptr) { __device__ inline void *__nv_cvta_local_to_generic_impl(size_t __ptr) { return (void *)(void __attribute__((address_space(5))) *)__ptr; } -__device__ inline uint32_t __nvvm_get_smem_pointer(void *__ptr) { +__device__ inline cuuint32_t __nvvm_get_smem_pointer(void *__ptr) { return __nv_cvta_generic_to_shared_impl(__ptr); } } // extern "C" |