add some simple utilities for manipulating lists of types (#2370)

NVIDIA · Sep 16, 2024 · 684cf8e · 684cf8e
1 parent cf21a40
commit 684cf8e
Show file tree

Hide file tree

Showing 19 changed files with 2,844 additions and 835 deletions.
diff --git a/cub/test/CMakeLists.txt b/cub/test/CMakeLists.txt
@@ -21,11 +21,6 @@ endif()
 include("${CUB_SOURCE_DIR}/cmake/CPM.cmake")
 CPMAddPackage("gh:catchorg/[email protected]")
 
-option(METAL_BUILD_DOC OFF)
-option(METAL_BUILD_EXAMPLES OFF)
-option(METAL_BUILD_TESTS OFF)
-CPMAddPackage("gh:brunocodutra/[email protected]")
-
 CPMAddPackage(
   NAME NVTX
   GITHUB_REPOSITORY NVIDIA/NVTX
@@ -224,7 +219,6 @@ function(cub_add_test target_name_var test_name test_src cub_target launcher_id)
         target_link_libraries(${config_c2run_target} PRIVATE
           ${cub_target}
           ${config_c2h_target}
-          Metal
           Catch2::Catch2)
         cub_clone_target_properties(${config_c2run_target} ${cub_target})
         cub_configure_cuda_target(${config_c2run_target} RDC ${cdp_val})
@@ -269,7 +263,6 @@ function(cub_add_test target_name_var test_name test_src cub_target launcher_id)
     target_link_libraries(${test_target} PRIVATE
       ${cub_target}
       ${config_c2h_target}
-      Metal
       Catch2::Catch2
     )
     cub_clone_target_properties(${test_target} ${cub_target})

diff --git a/cub/test/catch2_test_device_histogram.cu b/cub/test/catch2_test_device_histogram.cu
@@ -430,7 +430,7 @@ CUB_TEST("DeviceHistogram::Histogram* basic use", "[histogram][device]", types)
 
 // TODO(bgruber): float produces INFs in the HistogramRange test setup AND the HistogramEven implementation
 // This test covers int32 and int64 arithmetic for bin computation
-CUB_TEST("DeviceHistogram::Histogram* large levels", "[histogram][device]", metal::remove<types, float>)
+CUB_TEST("DeviceHistogram::Histogram* large levels", "[histogram][device]", c2h::remove<types, float>)
 {
   using sample_t             = c2h::get<0, TestType>;
   using level_t              = sample_t;

diff --git a/cub/test/catch2_test_helper.h b/cub/test/catch2_test_helper.h
@@ -36,14 +36,14 @@
 
 #include "cuda/std/__cccl/diagnostic.h"
 #include "test_util_vec.h"
-#include <metal.hpp>
 
 #if __CUDACC_VER_MAJOR__ == 11
 _CCCL_NV_DIAG_SUPPRESS(177) // catch2 may contain unused variableds
 #endif // nvcc-11
 
 #include <cuda/std/bit>
 #include <cuda/std/cmath>
+#include <cuda/std/type_traits>
 #include <cuda/std/utility>
 
 #include "catch2_main.cuh"
@@ -60,31 +60,34 @@ namespace c2h
 {
 
 template <typename... Ts>
-using type_list = metal::list<Ts...>;
+using type_list = ::cuda::std::__type_list<Ts...>;
 
 template <typename TypeList>
-using size = metal::size<TypeList>;
+using size = ::cuda::std::__type_list_size<TypeList>;
 
 template <std::size_t Index, typename TypeList>
-using get = metal::at<TypeList, metal::number<Index>>;
+using get = ::cuda::std::__type_at_c<Index, TypeList>;
 
 template <class... TypeLists>
-using cartesian_product = metal::cartesian<TypeLists...>;
+using cartesian_product = ::cuda::std::__type_cartesian_product<TypeLists...>;
 
 template <typename T, T... Ts>
-using enum_type_list = c2h::type_list<std::integral_constant<T, Ts>...>;
+using enum_type_list = ::cuda::std::__type_value_list<T, Ts...>;
 
 template <typename T0, typename T1>
-using pair = metal::pair<T0, T1>;
+using pair = ::cuda::std::__type_pair<T0, T1>;
 
 template <typename P>
-using first = metal::first<P>;
+using first = ::cuda::std::__type_pair_first<P>;
 
 template <typename P>
-using second = metal::second<P>;
+using second = ::cuda::std::__type_pair_second<P>;
 
-template <std::size_t start, std::size_t size, std::size_t stride = 1>
-using iota = metal::iota<metal::number<start>, metal::number<size>, metal::number<stride>>;
+template <std::size_t Start, std::size_t Size, std::size_t Stride = 1>
+using iota = ::cuda::std::__type_iota<std::size_t, Start, Size, Stride>;
+
+template <typename TypeList, typename T>
+using remove = ::cuda::std::__type_remove<TypeList, T>;
 
 } // namespace c2h
 

diff --git a/cub/test/catch2_test_iterator.cu b/cub/test/catch2_test_iterator.cu
@@ -43,7 +43,7 @@
 
 using scalar_types = c2h::type_list<std::int8_t, std::int16_t, std::int32_t, std::int64_t, float, double>;
 
-using types = metal::append<
+using types = ::cuda::std::__type_push_back<
   scalar_types,
   char2,
   short2,

diff --git a/libcudacxx/include/cuda/__cccl_config b/libcudacxx/include/cuda/__cccl_config
@@ -18,6 +18,7 @@
 #include <cuda/std/__cccl/exceptions.h> // IWYU pragma: export
 #include <cuda/std/__cccl/execution_space.h> // IWYU pragma: export
 #include <cuda/std/__cccl/extended_floating_point.h> // IWYU pragma: export
+#include <cuda/std/__cccl/preprocessor.h> // IWYU pragma: export
 #include <cuda/std/__cccl/ptx_isa.h> // IWYU pragma: export
 #include <cuda/std/__cccl/sequence_access.h> // IWYU pragma: export
 #include <cuda/std/__cccl/system_header.h> // IWYU pragma: export

diff --git a/libcudacxx/include/cuda/std/__cccl/compiler.h b/libcudacxx/include/cuda/std/__cccl/compiler.h
@@ -90,6 +90,9 @@
 #if !defined(_CCCL_CUDA_COMPILER) || (defined(_CCCL_CUDACC) && _CCCL_CUDACC_VER < 1108000)
 #  define _CCCL_CUDACC_BELOW_11_8
 #endif // defined(_CCCL_CUDACC) && _CCCL_CUDACC_VER < 1108000
+#if !defined(_CCCL_CUDA_COMPILER) || (defined(_CCCL_CUDACC) && _CCCL_CUDACC_VER < 1200000)
+#  define _CCCL_CUDACC_BELOW_12_0
+#endif // defined(_CCCL_CUDACC) && _CCCL_CUDACC_VER < 1203000
 #if !defined(_CCCL_CUDA_COMPILER) || (defined(_CCCL_CUDACC) && _CCCL_CUDACC_VER < 1202000)
 #  define _CCCL_CUDACC_BELOW_12_2
 #endif // defined(_CCCL_CUDACC) && _CCCL_CUDACC_VER < 1203000