[PATCH] D80464: [CUDA] Missing __syncthreads intrinsic in __clang_cuda_device_functions.h

[Boris Staletic via Phabricator via cfe-commits]

bstaletic created this revision.
bstaletic added a reviewer: aprantl.
Herald added subscribers: cfe-commits, yaxunl.
Herald added a project: clang.

Seems like the `__syncthreads` is missing from the `clang/lib/Headers/__clang_cuda_device_functions.h` file. To be honest, I don't know much about CUDA. This issue was noticed by a YouCompleteMe user who then made a pull request:

https://github.com/ycm-core/ycmd/pull/1438

I did not create any tests, because a similar patch did not include tests:

https://reviews.llvm.org/D43602


Repository:
  rG LLVM Github Monorepo

https://reviews.llvm.org/D80464

Files:
  clang/lib/Headers/__clang_cuda_device_functions.h


Index: clang/lib/Headers/__clang_cuda_device_functions.h
===================================================================
--- clang/lib/Headers/__clang_cuda_device_functions.h
+++ clang/lib/Headers/__clang_cuda_device_functions.h
@@ -519,6 +519,7 @@
   return __nv_fast_sincosf(__a, __s, __c);
 }
 __DEVICE__ float __sinf(float __a) { return __nv_fast_sinf(__a); }
+__DEVICE__ int __syncthreads(void) { return __nvvm_bar0(); }
 __DEVICE__ int __syncthreads_and(int __a) { return __nvvm_bar0_and(__a); }
 __DEVICE__ int __syncthreads_count(int __a) { return __nvvm_bar0_popc(__a); }
 __DEVICE__ int __syncthreads_or(int __a) { return __nvvm_bar0_or(__a); }


-------------- next part --------------
A non-text attachment was scrubbed...
Name: D80464.265817.patch
Type: text/x-patch
Size: 659 bytes
Desc: not available
URL: <http://lists.llvm.org/pipermail/cfe-commits/attachments/20200523/d303dbf3/attachment.bin>