[CUDA] add wrapper header for libc++'s __utlility/declval.h #148918
Merged
Conversation
This file contains hidden or bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
Since llvm#116709 more libc++ code relies on std::declval() and it broke some CUDA compilations. The new wrapper adds GPU-side overloads for the declval() helper functions which allows it to continue working when used from CUDA sources.
llvmbot
added
clang
|
@llvm/pr-subscribers-backend-x86 @llvm/pr-subscribers-clang Author: Artem Belevich (Artem-B) ChangesSince #116709 more libc++ code relies on std::declval() and it broke some CUDA compilations. The new wrapper adds GPU-side overloads for the declval() helper functions which allows it to continue working when used from CUDA sources. Full diff: https://github.com/llvm/llvm-project/pull/148918.diff 2 Files Affected:
diff --git a/clang/lib/Headers/CMakeLists.txt b/clang/lib/Headers/CMakeLists.txt
index c96d209c1fc0c..b4618fe4a46da 100644
--- a/clang/lib/Headers/CMakeLists.txt
+++ b/clang/lib/Headers/CMakeLists.txt
@@ -347,6 +347,10 @@ set(cuda_wrapper_bits_files
cuda_wrappers/bits/basic_string.tcc
)
+set(cuda_wrapper_utility_files
+ cuda_wrappers/__utility/declval.h
+)
+
set(ppc_wrapper_files
ppc_wrappers/mmintrin.h
ppc_wrappers/xmmintrin.h
@@ -443,8 +447,9 @@ endfunction(clang_generate_header)
# Copy header files from the source directory to the build directory
foreach( f ${files} ${cuda_wrapper_files} ${cuda_wrapper_bits_files}
- ${ppc_wrapper_files} ${openmp_wrapper_files} ${zos_wrapper_files} ${hlsl_files}
- ${llvm_libc_wrapper_files} ${llvm_offload_wrapper_files})
+ ${cuda_wrapper_utility_files} ${ppc_wrapper_files} ${openmp_wrapper_files}
+ ${zos_wrapper_files} ${hlsl_files} ${llvm_libc_wrapper_files}
+ ${llvm_offload_wrapper_files})
copy_header_to_output_dir(${CMAKE_CURRENT_SOURCE_DIR} ${f})
endforeach( f )
@@ -553,7 +558,7 @@ add_header_target("arm-common-resource-headers" "${arm_common_files};${arm_commo
# Architecture/platform specific targets
add_header_target("arm-resource-headers" "${arm_only_files};${arm_only_generated_files}")
add_header_target("aarch64-resource-headers" "${aarch64_only_files};${aarch64_only_generated_files}")
-add_header_target("cuda-resource-headers" "${cuda_files};${cuda_wrapper_files};${cuda_wrapper_bits_files}")
+add_header_target("cuda-resource-headers" "${cuda_files};${cuda_wrapper_files};${cuda_wrapper_bits_files};${cuda_wrapper_utility_files}")
add_header_target("hexagon-resource-headers" "${hexagon_files}")
add_header_target("hip-resource-headers" "${hip_files}")
add_header_target("loongarch-resource-headers" "${loongarch_files}")
@@ -600,6 +605,11 @@ install(
DESTINATION ${header_install_dir}/cuda_wrappers/bits
COMPONENT clang-resource-headers)
+install(
+ FILES ${cuda_wrapper_utility_files}
+ DESTINATION ${header_install_dir}/cuda_wrappers/__utility
+ COMPONENT clang-resource-headers)
+
install(
FILES ${ppc_wrapper_files}
DESTINATION ${header_install_dir}/ppc_wrappers
@@ -663,6 +673,12 @@ install(
EXCLUDE_FROM_ALL
COMPONENT cuda-resource-headers)
+install(
+ FILES ${cuda_wrapper_utility_files}
+ DESTINATION ${header_install_dir}/cuda_wrappers/__utility
+ EXCLUDE_FROM_ALL
+ COMPONENT cuda-resource-headers)
+
install(
FILES ${cuda_files}
DESTINATION ${header_install_dir}
diff --git a/clang/lib/Headers/cuda_wrappers/__utility/declval.h b/clang/lib/Headers/cuda_wrappers/__utility/declval.h
new file mode 100644
index 0000000000000..b16311e849fa4
--- /dev/null
+++ b/clang/lib/Headers/cuda_wrappers/__utility/declval.h
@@ -0,0 +1,31 @@
+#ifndef __CUDA_WRAPPERS_UTILITY_DECLVAL_H__
+#define __CUDA_WRAPPERS_UTILITY_DECLVAL_H__
+
+#include_next <__utility/declval.h>
+
+// The stuff below is the exact copy of the <__utility/declval.h>,
+// but with __device__ attribute applied to the functions, so it works on a GPU.
+
+_LIBCPP_BEGIN_NAMESPACE_STD
+
+// Suppress deprecation notice for volatile-qualified return type resulting
+// from volatile-qualified types _Tp.
+_LIBCPP_SUPPRESS_DEPRECATED_PUSH
+template <class _Tp>
+__attribute__((device))
+_Tp&& __declval(int);
+template <class _Tp>
+__attribute__((device))
+_Tp __declval(long);
+_LIBCPP_SUPPRESS_DEPRECATED_POP
+
+template <class _Tp>
+__attribute__((device))
+_LIBCPP_HIDE_FROM_ABI decltype(std::__declval<_Tp>(0)) declval() _NOEXCEPT {
+ static_assert(!__is_same(_Tp, _Tp),
+ "std::declval can only be used in an unevaluated context. "
+ "It's likely that your current usage is trying to extract a value from the function.");
+}
+
+_LIBCPP_END_NAMESPACE_STD
+#endif // __CUDA_WRAPPERS_UTILITY_DECLVAL_H__
|
|
✅ With the latest revision this PR passed the C/C++ code formatter. |
|
LGTM. I would like it if the new file were formatted, but I suppose it's intentional in order to be more of an exact copy of the libc++ file? |
|
LLVM Buildbot has detected a new failure on builder Full details are available at: https://lab.llvm.org/buildbot/#/builders/88/builds/13968 Here is the relevant piece of the build log for the reference |
|
LLVM Buildbot has detected a new failure on builder Full details are available at: https://lab.llvm.org/buildbot/#/builders/153/builds/38242 Here is the relevant piece of the build log for the reference |
Sign up for free
to join this conversation on GitHub.
Already have an account?
Sign in to comment
4 participants
Add this suggestion to a batch that can be applied as a single commit.
This suggestion is invalid because no changes were made to the code.
Suggestions cannot be applied while the pull request is closed.
Suggestions cannot be applied while viewing a subset of changes.
Only one suggestion per line can be applied in a batch.
Add this suggestion to a batch that can be applied as a single commit.
Applying suggestions on deleted lines is not supported.
You must change the existing code in this line in order to create a valid suggestion.
Outdated suggestions cannot be applied.
This suggestion has been applied or marked resolved.
Suggestions cannot be applied from pending reviews.
Suggestions cannot be applied on multi-line comments.
Suggestions cannot be applied while the pull request is queued to merge.
Suggestion cannot be applied right now. Please check back later.
Since #116709 more libc++ code relies on std::declval() and it broke some CUDA compilations.
The new wrapper adds GPU-side overloads for the declval() helper functions which allows it to continue working when used from CUDA sources.