Cleanup transitive includes of <cuda/std/functional> (#1253)

functional is currently pulling in a lot of unnecessary headers that significantly increase compile times. Fix this by including only necessary ones
NVIDIA · Jan 30, 2024 · 8db2804 · 8db2804
1 parent b7d4228
commit 8db2804
Show file tree

Hide file tree

Showing 15 changed files with 34 additions and 13 deletions.
diff --git a/cub/benchmarks/bench/histogram/histogram_common.cuh b/cub/benchmarks/bench/histogram/histogram_common.cuh
@@ -29,6 +29,8 @@
 
 #include <cub/device/device_histogram.cuh>
 
+#include <cuda/std/type_traits>
+
 #if !TUNE_BASE
 
 #if TUNE_LOAD == 0
@@ -37,7 +39,7 @@
 #define TUNE_LOAD_MODIFIER cub::LOAD_LDG
 #else // TUNE_LOAD == 2
 #define TUNE_LOAD_MODIFIER cub::LOAD_CA
-#endif // TUNE_LOAD 
+#endif // TUNE_LOAD
 
 #define TUNE_VEC_SIZE (1 << TUNE_VEC_SIZE_POW)
 
@@ -51,11 +53,11 @@ constexpr cub::BlockHistogramMemoryPreference MEM_PREFERENCE = cub::BLEND;
 
 #if TUNE_LOAD_ALGORITHM_ID == 0
 #define TUNE_LOAD_ALGORITHM cub::BLOCK_LOAD_DIRECT
-#elif TUNE_LOAD_ALGORITHM_ID == 1 
+#elif TUNE_LOAD_ALGORITHM_ID == 1
 #define TUNE_LOAD_ALGORITHM cub::BLOCK_LOAD_WARP_TRANSPOSE
-#else 
+#else
 #define TUNE_LOAD_ALGORITHM cub::BLOCK_LOAD_STRIPED
-#endif // TUNE_LOAD_ALGORITHM_ID 
+#endif // TUNE_LOAD_ALGORITHM_ID
 
 template <typename SampleT, int NUM_CHANNELS, int NUM_ACTIVE_CHANNELS>
 struct policy_hub_t

diff --git a/cub/cub/agent/agent_reduce.cuh b/cub/cub/agent/agent_reduce.cuh
@@ -53,7 +53,9 @@
 #include <cub/iterator/cache_modified_input_iterator.cuh>
 #include <cub/util_type.cuh>
 
+_CCCL_DIAG_SUPPRESS_DEPRECATED_PUSH
 #include <cuda/std/functional>
+_CCCL_DIAG_SUPPRESS_DEPRECATED_POP
 
 CUB_NAMESPACE_BEGIN
 

diff --git a/cub/cub/agent/agent_three_way_partition.cuh b/cub/cub/agent/agent_three_way_partition.cuh
@@ -48,6 +48,7 @@
 #include <cub/block/block_store.cuh>
 #include <cub/iterator/cache_modified_input_iterator.cuh>
 
+#include <cuda/std/type_traits>
 
 CUB_NAMESPACE_BEGIN
 

diff --git a/cub/cub/block/block_radix_rank.cuh b/cub/cub/block/block_radix_rank.cuh
@@ -47,9 +47,9 @@
 #include <cub/util_ptx.cuh>
 #include <cub/util_type.cuh>
 
-#include <cuda/std/type_traits>
 #include <cuda/std/cstdint>
-
+#include <cuda/std/limits>
+#include <cuda/std/type_traits>
 
 CUB_NAMESPACE_BEGIN
 

diff --git a/cub/cub/block/radix_rank_sort_operations.cuh b/cub/cub/block/radix_rank_sort_operations.cuh
@@ -50,6 +50,7 @@
 
 #include <thrust/type_traits/integer_sequence.h>
 
+#include <cuda/std/cstdint>
 #include <cuda/std/tuple>
 #include <cuda/std/type_traits>
 

diff --git a/cub/cub/detail/type_traits.cuh b/cub/cub/detail/type_traits.cuh
@@ -45,9 +45,11 @@
 #include <cub/util_cpp_dialect.cuh>
 #include <cub/util_namespace.cuh>
 
+_CCCL_DIAG_SUPPRESS_DEPRECATED_PUSH
+#include <cuda/std/functional>
+_CCCL_DIAG_SUPPRESS_DEPRECATED_POP
 #include <cuda/std/type_traits>
 
-
 CUB_NAMESPACE_BEGIN
 namespace detail {
 

diff --git a/cub/cub/detail/uninitialized_copy.cuh b/cub/cub/detail/uninitialized_copy.cuh
@@ -37,8 +37,8 @@
 #  pragma system_header
 #endif // no system header
 
-
 #include <cuda/std/type_traits>
+#include <cuda/std/utility>
 
 CUB_NAMESPACE_BEGIN
 

diff --git a/cub/cub/device/device_select.cuh b/cub/cub/device/device_select.cuh
@@ -47,6 +47,8 @@
 #include <cub/device/dispatch/dispatch_unique_by_key.cuh>
 #include <cub/util_deprecated.cuh>
 
+#include <cuda/std/type_traits>
+
 #include <iterator>
 #include <stdio.h>
 

diff --git a/cub/cub/device/dispatch/dispatch_reduce.cuh b/cub/cub/device/dispatch/dispatch_reduce.cuh
@@ -58,7 +58,9 @@
 
 #include <iterator>
 
+_CCCL_DIAG_SUPPRESS_DEPRECATED_PUSH
 #include <cuda/std/functional>
+_CCCL_DIAG_SUPPRESS_DEPRECATED_POP
 
 #include <stdio.h>
 

diff --git a/cub/cub/util_device.cuh b/cub/cub/util_device.cuh
@@ -56,6 +56,7 @@
 
 #include <cuda/discard_memory>
 #include <cuda/std/utility>
+#include <cuda/std/type_traits>
 
 #include <nv/target>
 

diff --git a/cub/cub/util_ptx.cuh b/cub/cub/util_ptx.cuh
@@ -465,6 +465,7 @@ unsigned int WarpMask(unsigned int warp_id)
   {
     member_mask <<= warp_id * LOGICAL_WARP_THREADS;
   }
+  (void)warp_id;
 
   return member_mask;
 }

diff --git a/cub/cub/util_type.cuh b/cub/cub/util_type.cuh
@@ -68,6 +68,8 @@
 #include <cub/detail/uninitialized_copy.cuh>
 
 #include <cuda/std/type_traits>
+#include <cuda/std/cstdint>
+#include <cuda/std/limits>
 
 #if !defined(_LIBCUDACXX_COMPILER_NVRTC)
 #  include <iterator>

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/functional b/libcudacxx/include/cuda/std/detail/libcxx/include/functional
@@ -505,7 +505,6 @@ POLICY:  For non-variadic implementations, the number of arguments is limited
 #endif // _LIBCUDACXX_COMPILER_NVRTC
 #endif // __cuda_std__
 
-
 #include "__functional_base"
 #include "__functional/binary_function.h"
 #include "__functional/binary_negate.h"
@@ -535,10 +534,7 @@ POLICY:  For non-variadic implementations, the number of arguments is limited
 #include "__functional/unwrap_ref.h"
 #include "__functional/weak_result_type.h"
 
-#include "chrono"
-#include "climits"
-#include "iterator"
-#include "type_traits"
+#include "iosfwd" // for forward declarations of vector and string.
 #include "version"
 
 #if defined(_CCCL_IMPLICIT_SYSTEM_HEADER_GCC)

diff --git a/libcudacxx/test/libcudacxx/std/utilities/meta/meta.rel/is_nothrow_invocable.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/meta/meta.rel/is_nothrow_invocable.pass.cpp
@@ -95,12 +95,14 @@ int main(int, char**) {
     static_assert(!cuda::std::is_nothrow_invocable<const int&>::value, "");
     static_assert(!cuda::std::is_nothrow_invocable<int&&>::value, "");
 
+#ifdef _LIBCUDACXX_HAS_VECTOR
     static_assert(!cuda::std::is_nothrow_invocable<int, cuda::std::vector<int> >::value,
                   "");
     static_assert(!cuda::std::is_nothrow_invocable<int, cuda::std::vector<int*> >::value,
                   "");
     static_assert(!cuda::std::is_nothrow_invocable<int, cuda::std::vector<int**> >::value,
                   "");
+#endif // _LIBCUDACXX_HAS_VECTOR
 
 #if TEST_STD_VER >= 2017
     static_assert(!cuda::std::is_nothrow_invocable<AbominableFunc>::value, "");
@@ -138,12 +140,14 @@ int main(int, char**) {
     static_assert(!cuda::std::is_nothrow_invocable_r<int, const int&>::value, "");
     static_assert(!cuda::std::is_nothrow_invocable_r<int, int&&>::value, "");
 
+#ifdef _LIBCUDACXX_HAS_VECTOR
     static_assert(!cuda::std::is_nothrow_invocable_r<int, cuda::std::vector<int> >::value,
                   "");
     static_assert(!cuda::std::is_nothrow_invocable_r<int, cuda::std::vector<int*> >::value,
                   "");
     static_assert(!cuda::std::is_nothrow_invocable_r<int, cuda::std::vector<int**> >::value,
                   "");
+#endif // _LIBCUDACXX_HAS_VECTOR
 #if TEST_STD_VER >= 2017
     static_assert(!cuda::std::is_nothrow_invocable_r<void, AbominableFunc>::value,
                   "");

diff --git a/thrust/thrust/detail/allocator/allocator_traits.h b/thrust/thrust/detail/allocator/allocator_traits.h
@@ -29,6 +29,7 @@
 #elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_MSVC)
 #  pragma system_header
 #endif // no system header
+
 #include <thrust/detail/type_traits/pointer_traits.h>
 #include <thrust/detail/type_traits/has_nested_type.h>
 #include <thrust/detail/type_traits/has_member_function.h>
@@ -80,6 +81,8 @@ template<typename Alloc, typename U>
   typedef thrust::detail::integral_constant<bool, value> type;
 };
 
+_CCCL_DIAG_SUPPRESS_DEPRECATED_PUSH
+
 // The following fields of std::allocator have been deprecated (since C++17).
 // There's no way to detect it other than explicit specialization.
 #if _CCCL_STD_VER >= 2017
@@ -186,6 +189,8 @@ template<typename Alloc>
   static const bool value = type::value;
 };
 
+_CCCL_DIAG_SUPPRESS_DEPRECATED_POP
+
 template<class Alloc, class U, bool = has_rebind<Alloc, U>::value>
   struct rebind_alloc
 {