Author: Artem Belevich Date: 2022-11-17T11:38:16-08:00 New Revision: 817340569bf98b696329c53508a0d87cc0daec25
URL: https://github.com/llvm/llvm-project/commit/817340569bf98b696329c53508a0d87cc0daec25 DIFF: https://github.com/llvm/llvm-project/commit/817340569bf98b696329c53508a0d87cc0daec25.diff LOG: [CUDA] make use of deprecated texture API conditional on CUDA version. Added: Modified: clang/lib/Headers/__clang_cuda_texture_intrinsics.h Removed: ################################################################################ diff --git a/clang/lib/Headers/__clang_cuda_texture_intrinsics.h b/clang/lib/Headers/__clang_cuda_texture_intrinsics.h index 3c0f0026f1f02..a71952211237b 100644 --- a/clang/lib/Headers/__clang_cuda_texture_intrinsics.h +++ b/clang/lib/Headers/__clang_cuda_texture_intrinsics.h @@ -666,6 +666,7 @@ __device__ static void __tex_fetch(__T *__ptr, cudaTextureObject_t __handle, __tex_fetch_v4<__op>::template __run<__FetchT>(__handle, __args...)); } +#if CUDA_VERSION < 12000 // texture<> objects get magically converted into a texture reference. However, // there's no way to convert them to cudaTextureObject_t on C++ level. So, we // cheat a bit and use inline assembly to do it. It costs us an extra register @@ -713,6 +714,7 @@ __tex_fetch(__DataT *, __RetT *__ptr, __tex_fetch_v4<__op>::template __run<__FetchT>( __tex_handle_to_obj(__handle), __args...)); } +#endif // CUDA_VERSION } // namespace __cuda_tex } // namespace #pragma pop_macro("__ASM_OUT") _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
