Update to CCCL 2.2.0.

ci/build_cpp.sh

@@ -9,12 +9,15 @@ export CMAKE_GENERATOR=Ninja
 
 rapids-print-env
 
+LIBRMM_CHANNEL=$(rapids-get-pr-conda-artifact rmm 1404 cpp)
+
 version=$(rapids-generate-version)
 
 rapids-logger "Begin cpp build"
 
 # With boa installed conda build forward to boa
 RAPIDS_PACKAGE_VERSION=${version} rapids-conda-retry mambabuild \
+    --channel "${LIBRMM_CHANNEL}" \
     conda/recipes/libcudf
 
 rapids-upload-conda-to-s3 cpp

ci/build_python.sh

@@ -21,30 +21,40 @@ done
 rapids-logger "Begin py build"
 
 CPP_CHANNEL=$(rapids-download-conda-from-s3 cpp)
+LIBRMM_CHANNEL=$(rapids-get-pr-conda-artifact rmm 1404 cpp)
+RMM_CHANNEL=$(rapids-get-pr-conda-artifact rmm 1404 python)
 
 # TODO: Remove `--no-test` flag once importing on a CPU
 # node works correctly
 # With boa installed conda build forwards to the boa builder
 RAPIDS_PACKAGE_VERSION=${version} rapids-conda-retry mambabuild \
   --no-test \
   --channel "${CPP_CHANNEL}" \
+  --channel "${LIBRMM_CHANNEL}" \
+  --channel "${RMM_CHANNEL}" \
   conda/recipes/cudf
 
 RAPIDS_PACKAGE_VERSION=${version} rapids-conda-retry mambabuild \
   --no-test \
   --channel "${CPP_CHANNEL}" \
+  --channel "${LIBRMM_CHANNEL}" \
+  --channel "${RMM_CHANNEL}" \
   --channel "${RAPIDS_CONDA_BLD_OUTPUT_DIR}" \
   conda/recipes/dask-cudf
 
 RAPIDS_PACKAGE_VERSION=${version} rapids-conda-retry mambabuild \
   --no-test \
   --channel "${CPP_CHANNEL}" \
+  --channel "${LIBRMM_CHANNEL}" \
+  --channel "${RMM_CHANNEL}" \
   --channel "${RAPIDS_CONDA_BLD_OUTPUT_DIR}" \
   conda/recipes/cudf_kafka
 
 RAPIDS_PACKAGE_VERSION=${version} rapids-conda-retry mambabuild \
   --no-test \
   --channel "${CPP_CHANNEL}" \
+  --channel "${LIBRMM_CHANNEL}" \
+  --channel "${RMM_CHANNEL}" \
   --channel "${RAPIDS_CONDA_BLD_OUTPUT_DIR}" \
   conda/recipes/custreamz
 

ci/test_cpp_common.sh

@@ -22,6 +22,8 @@ conda activate test
 set -u
 
 CPP_CHANNEL=$(rapids-download-conda-from-s3 cpp)
+LIBRMM_CHANNEL=$(rapids-get-pr-conda-artifact rmm 1404 cpp)
+
 RESULTS_DIR=${RAPIDS_TESTS_DIR:-"$(mktemp -d)"}
 RAPIDS_TESTS_DIR=${RAPIDS_TESTS_DIR:-"${RESULTS_DIR}/test-results"}/
 mkdir -p "${RAPIDS_TESTS_DIR}"
@@ -30,6 +32,7 @@ rapids-print-env
 
 rapids-mamba-retry install \
   --channel "${CPP_CHANNEL}" \
+  --channel "${LIBRMM_CHANNEL}" \
   libcudf libcudf_kafka libcudf-tests
 
 rapids-logger "Check GPU usage"

ci/test_java.sh

@@ -25,11 +25,14 @@ set -u
 
 rapids-print-env
 
+LIBRMM_CHANNEL=$(rapids-get-pr-conda-artifact rmm 1404 cpp)
+
 rapids-logger "Downloading artifacts from previous jobs"
 CPP_CHANNEL=$(rapids-download-conda-from-s3 cpp)
 
 rapids-mamba-retry install \
   --channel "${CPP_CHANNEL}" \
+  --channel "${LIBRMM_CHANNEL}" \
   libcudf
 
 rapids-logger "Check GPU usage"

ci/test_python_common.sh

@@ -26,6 +26,8 @@ set -u
 rapids-logger "Downloading artifacts from previous jobs"
 CPP_CHANNEL=$(rapids-download-conda-from-s3 cpp)
 PYTHON_CHANNEL=$(rapids-download-conda-from-s3 python)
+LIBRMM_CHANNEL=$(rapids-get-pr-conda-artifact rmm 1404 cpp)
+RMM_CHANNEL=$(rapids-get-pr-conda-artifact rmm 1404 python)
 
 RESULTS_DIR=${RAPIDS_TESTS_DIR:-"$(mktemp -d)"}
 RAPIDS_TESTS_DIR=${RAPIDS_TESTS_DIR:-"${RESULTS_DIR}/test-results"}/
@@ -37,4 +39,6 @@ rapids-print-env
 rapids-mamba-retry install \
   --channel "${CPP_CHANNEL}" \
   --channel "${PYTHON_CHANNEL}" \
+  --channel "${LIBRMM_CHANNEL}" \
+  --channel "${RMM_CHANNEL}" \
   cudf libcudf

cpp/CMakeLists.txt

@@ -183,17 +183,15 @@ rapids_cpm_init()
 include(cmake/thirdparty/get_jitify.cmake)
 # find nvCOMP
 include(cmake/thirdparty/get_nvcomp.cmake)
-# find thrust/cub
-include(cmake/thirdparty/get_thrust.cmake)
+# find CCCL before rmm so that we get cudf's patched version of CCCL
+include(cmake/thirdparty/get_cccl.cmake)
 # find rmm
 include(cmake/thirdparty/get_rmm.cmake)
 # find arrow
 include(cmake/thirdparty/get_arrow.cmake)
 # find dlpack
 include(cmake/thirdparty/get_dlpack.cmake)
-# find libcu++
-include(cmake/thirdparty/get_libcudacxx.cmake)
-# find cuCollections Should come after including thrust and libcudacxx
+# find cuCollections, should come after including CCCL
 include(cmake/thirdparty/get_cucollections.cmake)
 # find or install GoogleTest
 if(CUDF_BUILD_TESTUTIL)
@@ -758,7 +756,7 @@ add_dependencies(cudf jitify_preprocess_run)
 # Specify the target module library dependencies
 target_link_libraries(
   cudf
-  PUBLIC ${ARROW_LIBRARIES} libcudacxx::libcudacxx cudf::Thrust rmm::rmm
+  PUBLIC ${ARROW_LIBRARIES} CCCL::CCCL rmm::rmm
   PRIVATE cuco::cuco ZLIB::ZLIB nvcomp::nvcomp kvikio::kvikio
           $<TARGET_NAME_IF_EXISTS:cuFile_interface>
 )
@@ -1009,14 +1007,6 @@ following IMPORTED GLOBAL  targets:
     ]=]
 )
 
-set(common_code_string
-    [=[
-if(NOT TARGET cudf::Thrust)
-  thrust_create_target(cudf::Thrust FROM_OPTIONS)
-endif()
-]=]
-)
-
 if(CUDF_ENABLE_ARROW_PARQUET)
   string(
     APPEND
@@ -1040,7 +1030,6 @@ if(testing IN_LIST cudf_FIND_COMPONENTS)
 endif()
 ]=]
 )
-string(APPEND install_code_string "${common_code_string}")
 
 rapids_export(
   INSTALL cudf
@@ -1064,8 +1053,6 @@ endif()
 ]=]
 )
 
-string(APPEND build_code_string "${common_code_string}")
-
 rapids_export(
   BUILD cudf
   EXPORT_SET cudf-exports ${_components_export_string}

cpp/cmake/thirdparty/get_thrust.cmake → cpp/cmake/thirdparty/get_cccl.cmake

@@ -1,5 +1,5 @@
 # =============================================================================
-# Copyright (c) 2020-2023, NVIDIA CORPORATION.
+# Copyright (c) 2023, NVIDIA CORPORATION.
 #
 # Licensed under the Apache License, Version 2.0 (the "License"); you may not use this file except
 # in compliance with the License. You may obtain a copy of the License at
@@ -12,34 +12,30 @@
 # the License.
 # =============================================================================
 
-# This function finds thrust and sets any additional necessary environment variables.
-function(find_and_configure_thrust)
+# This function finds cccl and sets any additional necessary environment variables.
+function(find_and_configure_cccl)
 
-  include(${rapids-cmake-dir}/cpm/thrust.cmake)
+  include(${rapids-cmake-dir}/cpm/cccl.cmake)
   include(${rapids-cmake-dir}/cpm/package_override.cmake)
 
   set(cudf_patch_dir "${CMAKE_CURRENT_FUNCTION_LIST_DIR}/patches")
-  rapids_cpm_package_override("${cudf_patch_dir}/thrust_override.json")
+  rapids_cpm_package_override("${cudf_patch_dir}/cccl_override.json")
 
-  # Make sure we install thrust into the `include/libcudf` subdirectory instead of the default
+  # Make sure we install cccl into the `include/libcudf` subdirectory instead of the default
   include(GNUInstallDirs)
   set(CMAKE_INSTALL_INCLUDEDIR "${CMAKE_INSTALL_INCLUDEDIR}/libcudf")
   set(CMAKE_INSTALL_LIBDIR "${CMAKE_INSTALL_INCLUDEDIR}/lib")
 
-  # Find or install Thrust with our custom set of patches
-  rapids_cpm_thrust(
-    NAMESPACE cudf
-    BUILD_EXPORT_SET cudf-exports
-    INSTALL_EXPORT_SET cudf-exports
-  )
+  # Find or install CCCL with our custom set of patches
+  rapids_cpm_cccl(BUILD_EXPORT_SET cudf-exports INSTALL_EXPORT_SET cudf-exports)
 
-  # Store where CMake can find our custom Thrust install
+  # Store where CMake can find our custom CCCL install
   include("${rapids-cmake-dir}/export/find_package_root.cmake")
   rapids_export_find_package_root(
-    INSTALL Thrust [=[${CMAKE_CURRENT_LIST_DIR}/../../../include/libcudf/lib/rapids/cmake/thrust]=]
+    INSTALL CCCL [=[${CMAKE_CURRENT_LIST_DIR}/../../../include/libcudf/lib/rapids/cmake/cccl]=]
     EXPORT_SET cudf-exports
-    CONDITION Thrust_SOURCE_DIR
+    CONDITION CCCL_SOURCE_DIR
   )
 endfunction()
 
-find_and_configure_thrust()
+find_and_configure_cccl()

cpp/cmake/thirdparty/patches/thrust_override.json → cpp/cmake/thirdparty/patches/cccl_override.json

@@ -1,17 +1,12 @@
 
 {
   "packages" : {
-    "Thrust" : {
+    "CCCL" : {
       "patches" : [
         {
-          "file" : "Thrust/install_rules.diff",
-          "issue" : "Thrust 1.X installs incorrect files [https://github.com/NVIDIA/thrust/issues/1790]",
-          "fixed_in" : "2.0.0"
-        },
-        {
-          "file" : "${current_json_dir}/thrust_transform_iter_with_reduce_by_key.diff",
-          "issue" : "Support transform_output_iterator as output of reduce by key [https://github.com/NVIDIA/thrust/pull/1805]",
-          "fixed_in" : "2.1"
+          "file" : "cccl/bug_fixes.diff",
+          "issue" : "CCCL installs header-search.cmake files in nondeterministic order and has a typo in checking target creation that leads to duplicates",
+          "fixed_in" : "2.3"
         },
         {
           "file" : "${current_json_dir}/thrust_disable_64bit_dispatching.diff",
@@ -27,11 +22,6 @@
           "file" : "${current_json_dir}/thrust_faster_scan_compile_times.diff",
           "issue" : "Improve Thrust scan compile times by reducing the number of kernels generated [https://github.com/rapidsai/cudf/pull/8183]",
           "fixed_in" : ""
-        },
-        {
-          "file" : "${current_json_dir}/cub_segmented_sort_with_bool_key.diff",
-          "issue" : "Fix an error in CUB DeviceSegmentedSort when the keys are bool type [https://github.com/NVIDIA/cub/issues/594]",
-          "fixed_in" : "2.1"
         }
       ]
     }

cpp/cmake/thirdparty/patches/thrust_disable_64bit_dispatching.diff

@@ -1,29 +1,25 @@
-diff --git a/thrust/system/cuda/detail/dispatch.h b/thrust/system/cuda/detail/dispatch.h
-index d0e3f94..76774b0 100644
---- a/thrust/system/cuda/detail/dispatch.h
-+++ b/thrust/system/cuda/detail/dispatch.h
-@@ -32,9 +32,8 @@
+diff --git a/thrust/thrust/system/cuda/detail/dispatch.h b/thrust/thrust/system/cuda/detail/dispatch.h
+index d0e3f94ec..5c32a9c60 100644
+--- a/thrust/thrust/system/cuda/detail/dispatch.h
++++ b/thrust/thrust/system/cuda/detail/dispatch.h
+@@ -32,8 +32,7 @@
          status = call arguments; \
      } \
      else { \
 -        auto THRUST_PP_CAT2(count, _fixed) = static_cast<thrust::detail::int64_t>(count); \
 -        status = call arguments; \
--    }
-+       throw std::runtime_error("THRUST_INDEX_TYPE_DISPATCH 64-bit count is unsupported in libcudf"); \
-+    }
++        throw std::runtime_error("THRUST_INDEX_TYPE_DISPATCH 64-bit count is unsupported in libcudf"); \
+     }
 
  /**
-  * Dispatch between 32-bit and 64-bit index based versions of the same algorithm
-@@ -52,10 +51,8 @@
+@@ -52,9 +51,7 @@
          status = call arguments; \
      } \
      else { \
 -        auto THRUST_PP_CAT2(count1, _fixed) = static_cast<thrust::detail::int64_t>(count1); \
 -        auto THRUST_PP_CAT2(count2, _fixed) = static_cast<thrust::detail::int64_t>(count2); \
 -        status = call arguments; \
--    }
-+       throw std::runtime_error("THRUST_DOUBLE_INDEX_TYPE_DISPATCH 64-bit count is unsupported in libcudf"); \
-+    }
++        throw std::runtime_error("THRUST_DOUBLE_INDEX_TYPE_DISPATCH 64-bit count is unsupported in libcudf"); \
+     }
  /**
   * Dispatch between 32-bit and 64-bit index based versions of the same algorithm
-  * implementation. This version allows using different token sequences for callables