Merge branch 'main' into to_array

Author: galipremsagar

.agents/skills/build-test-cudf-java/SKILL.md

@@ -51,7 +51,7 @@ export MAVEN_OPTS="--add-opens java.base/java.lang=ALL-UNNAMED --add-opens java.
 Export `MVN_COMMON_OPTS` to match the CI build configuration in `java/ci/build-in-docker.sh`. For example:
 
 ```bash
-export MVN_COMMON_OPTS="-DCUDF_CPP_BUILD_DIR=$CUDF_CPP_BUILD_DIR -DBUILD_SHARED_LIBS=OFF -DCUDF_USE_PER_THREAD_DEFAULT_STREAM=ON -DCUDA_STATIC_CUFILE=ON -DCUDA_STATIC_RUNTIME=ON -DCUDF_JNI_LIBCUDF_STATIC=ON"
+export MVN_COMMON_OPTS="-DCUDF_CPP_BUILD_DIR=$CUDF_CPP_BUILD_DIR -DBUILD_SHARED_LIBS=OFF -DCUDF_USE_PER_THREAD_DEFAULT_STREAM=ON -DCUDA_STATIC_CUFILE=ON -DCUDF_JNI_LIBCUDF_STATIC=ON"
 ```
 
 ## Building cudf-java

.devcontainer/Dockerfile

@@ -37,7 +37,8 @@ ENV LIBCUDF_KERNEL_CACHE_PATH="/home/coder/cudf/cpp/build/${PYTHON_PACKAGE_MANAG
 ###
 # sccache configuration
 ###
-ENV AWS_ROLE_ARN="arn:aws:iam::279114543810:role/nv-gha-token-sccache-devs"
+ENV AWS_IDP_URL="https://token.rapids.nvidia.com"
+ENV AWS_ROLE_ARN="arn:aws:iam::279114543810:role/rapids-token-sccache-devs"
 ENV SCCACHE_REGION="us-east-2"
 ENV SCCACHE_BUCKET="rapids-sccache-devs"
 ENV SCCACHE_S3_USE_PREPROCESSOR_CACHE_MODE=true

conda/recipes/cudf/recipe.yaml

@@ -98,7 +98,6 @@ requirements:
     - pylibcudf =${{ version }}
     - ${{ pin_compatible("rmm", upper_bound="x.x") }}
     - fsspec >=0.6.0
-    - cuda-cudart
     - if: cuda_major == "12"
       then: cuda-python >=12.9.2,<13.0
       else: cuda-python >=13.0.1,<14.0

cpp/CMakeLists.txt

@@ -58,9 +58,6 @@ option(CUDA_ENABLE_LINEINFO
 )
 option(CUDA_WARNINGS_AS_ERRORS "Enable -Werror=all-warnings for all CUDA compilation" ON)
 
-# cudart can be statically linked or dynamically linked. The python ecosystem wants dynamic linking
-option(CUDA_STATIC_RUNTIME "Statically link the CUDA runtime" OFF)
-
 set(DEFAULT_CUDF_BUILD_STREAMS_TEST_UTIL ON)
 
 if(NOT BUILD_SHARED_LIBS)
@@ -97,7 +94,6 @@ message(
   VERBOSE
   "CUDF: Enable the -lineinfo option for nvcc (useful for cuda-memcheck / profiler): ${CUDA_ENABLE_LINEINFO}"
 )
-message(VERBOSE "CUDF: Statically link the CUDA runtime: ${CUDA_STATIC_RUNTIME}")
 message(VERBOSE
         "CUDF: Build with remote IO (e.g. AWS S3) support through KvikIO: ${CUDF_KVIKIO_REMOTE_IO}"
 )
@@ -1012,7 +1008,7 @@ if(TARGET conda_env)
   target_link_libraries(cudf PRIVATE conda_env)
 endif()
 
-rapids_cuda_set_runtime(cudf USE_STATIC ${CUDA_STATIC_RUNTIME})
+rapids_cuda_set_runtime(cudf USE_STATIC ON)
 
 file(
   WRITE "${CUDF_BINARY_DIR}/fatbin.ld"
@@ -1059,7 +1055,7 @@ if(CUDF_BUILD_TESTUTIL)
     PUBLIC cudf
     PRIVATE $<TARGET_NAME_IF_EXISTS:conda_env>
   )
-  rapids_cuda_set_runtime(cudftest_default_stream USE_STATIC ${CUDA_STATIC_RUNTIME})
+  rapids_cuda_set_runtime(cudftest_default_stream USE_STATIC ON)
 
   add_library(cudf::cudftest_default_stream ALIAS cudftest_default_stream)
 
@@ -1090,7 +1086,7 @@ if(CUDF_BUILD_TESTUTIL)
     cudftestutil INTERFACE "$<BUILD_INTERFACE:${CUDF_SOURCE_DIR}>"
                            "$<BUILD_INTERFACE:${CUDF_SOURCE_DIR}/src>"
   )
-  rapids_cuda_set_runtime(cudftestutil USE_STATIC ${CUDA_STATIC_RUNTIME})
+  rapids_cuda_set_runtime(cudftestutil USE_STATIC ON)
   add_library(cudf::cudftestutil ALIAS cudftestutil)
 
   add_library(cudftestutil_impl INTERFACE)
@@ -1151,7 +1147,7 @@ if(CUDF_BUILD_STREAMS_TEST_UTIL)
     endif()
 
     set(sanitizer_relative_genex
-        "$<PATH:RELATIVE_PATH,$<TARGET_FILE_DIR:CUDA::sanitizer>,$<TARGET_FILE_DIR:CUDA::cudart>>"
+        "$<PATH:RELATIVE_PATH,$<TARGET_FILE_DIR:CUDA::sanitizer>,$<TARGET_FILE_DIR:CUDA::cudart_static>>"
     )
     set_target_properties(
       ${_tgt}
@@ -1166,9 +1162,9 @@ if(CUDF_BUILD_STREAMS_TEST_UTIL)
       ${_tgt} PRIVATE "$<BUILD_INTERFACE:$<$<COMPILE_LANGUAGE:CXX>:${CUDF_CXX_FLAGS}>>"
     )
     target_include_directories(${_tgt} PRIVATE "$<BUILD_INTERFACE:${CUDF_SOURCE_DIR}/include>")
-    target_link_libraries(${_tgt} PUBLIC CUDA::cudart rmm::rmm CUDA::sanitizer)
+    target_link_libraries(${_tgt} PUBLIC CUDA::cudart_static rmm::rmm CUDA::sanitizer)
 
-    rapids_cuda_set_runtime(${_tgt} USE_STATIC ${CUDA_STATIC_RUNTIME})
+    rapids_cuda_set_runtime(${_tgt} USE_STATIC ON)
     add_library(cudf::${_tgt} ALIAS ${_tgt})
 
     if("${_mode}" STREQUAL "testing")

cpp/benchmarks/join/distinct_join.cu

@@ -18,11 +18,11 @@ void nvbench_distinct_inner_join(nvbench::state& state,
 {
   auto dtypes = cycle_dtypes(get_type_or_group(static_cast<int32_t>(DataType)), num_keys);
 
-  auto join = [](cudf::table_view const& probe_input,
-                 cudf::table_view const& build_input,
+  auto join = [](cudf::table_view const& left_input,
+                 cudf::table_view const& right_input,
                  cudf::null_equality compare_nulls) {
-    auto hj_obj = cudf::distinct_hash_join{build_input, compare_nulls, LOAD_FACTOR};
-    return hj_obj.inner_join(probe_input);
+    auto hj_obj = cudf::distinct_hash_join{right_input, compare_nulls, LOAD_FACTOR};
+    return hj_obj.inner_join(left_input);
   };
 
   BM_join<Nullable, join_t::HASH, NullEquality>(state, dtypes, join);
@@ -36,11 +36,11 @@ void nvbench_distinct_left_join(nvbench::state& state,
 {
   auto dtypes = cycle_dtypes(get_type_or_group(static_cast<int32_t>(DataType)), num_keys);
 
-  auto join = [](cudf::table_view const& probe_input,
-                 cudf::table_view const& build_input,
+  auto join = [](cudf::table_view const& left_input,
+                 cudf::table_view const& right_input,
                  cudf::null_equality compare_nulls) {
-    auto hj_obj = cudf::distinct_hash_join{build_input, compare_nulls, LOAD_FACTOR};
-    return hj_obj.left_join(probe_input);
+    auto hj_obj = cudf::distinct_hash_join{right_input, compare_nulls, LOAD_FACTOR};
+    return hj_obj.left_join(left_input);
   };
 
   BM_join<Nullable, join_t::HASH, NullEquality>(state, dtypes, join);

cpp/cmake/Modules/JitifyPreprocessKernels.cmake

@@ -9,7 +9,7 @@
 add_executable(jitify_preprocess "${JITIFY_INCLUDE_DIR}/jitify2_preprocess.cpp")
 
 target_compile_definitions(jitify_preprocess PRIVATE "_FILE_OFFSET_BITS=64")
-rapids_cuda_set_runtime(jitify_preprocess USE_STATIC ${CUDA_STATIC_RUNTIME})
+rapids_cuda_set_runtime(jitify_preprocess USE_STATIC ON)
 target_link_libraries(jitify_preprocess PUBLIC ${CMAKE_DL_LIBS})
 
 # Take a list of files to JIT-compile and run them through jitify_preprocess.

cpp/include/cudf/detail/join/distinct_hash_join.cuh

@@ -1,5 +1,5 @@
 /*
- * SPDX-FileCopyrightText: Copyright (c) 2024-2025, NVIDIA CORPORATION.
+ * SPDX-FileCopyrightText: Copyright (c) 2024-2026, NVIDIA CORPORATION.
  * SPDX-License-Identifier: Apache-2.0
  */
 #pragma once
@@ -25,14 +25,14 @@ using cudf::detail::row::lhs_index_type;
 using cudf::detail::row::rhs_index_type;
 
 /**
- * @brief A custom comparator used for the build table insertion
+ * @brief A custom comparator used for the right table insertion
  */
 struct always_not_equal {
   __device__ constexpr bool operator()(
     cuco::pair<hash_value_type, rhs_index_type> const&,
     cuco::pair<hash_value_type, rhs_index_type> const&) const noexcept
   {
-    // All build table keys are distinct thus `false` no matter what
+    // All right table keys are distinct thus `false` no matter what
     return false;
   }
 };
@@ -76,11 +76,11 @@ struct primitive_comparator_adapter {
 };
 
 /**
- * @brief Distinct hash join that builds hash table in creation and probes results in subsequent
- * `*_join` member functions.
+ * @brief Distinct hash join that builds a hash table with the right table on construction and
+ * probes results in subsequent `*_join` member functions.
  *
- * This class enables the distinct hash join scheme that builds hash table once, and probes as many
- * times as needed (possibly in parallel).
+ * This class enables the distinct hash join scheme that builds with the right table once and
+ * probes with many left tables (possibly in parallel).
  */
 class distinct_hash_join {
  public:
@@ -104,15 +104,15 @@ class distinct_hash_join {
   };
 
   /**
-   * @brief Constructor that internally builds the hash table based on the given `build` table.
+   * @brief Constructor that internally builds the hash table from the given `right` table.
    *
-   * @throw cudf::logic_error if the number of columns in `build` table is 0.
+   * @throw cudf::logic_error if the number of columns in `right` table is 0.
    *
-   * @param build The build table, from which the hash table is built
+   * @param right The right table, from which the hash table is built
    * @param compare_nulls Controls whether null join-key values should match or not.
    * @param stream CUDA stream used for device memory operations and kernel launches.
    */
-  distinct_hash_join(cudf::table_view const& build,
+  distinct_hash_join(cudf::table_view const& right,
                      cudf::null_equality compare_nulls,
                      rmm::cuda_stream_view stream);
 
@@ -121,7 +121,7 @@ class distinct_hash_join {
    *
    * @param load_factor The hash table occupancy ratio in (0,1]. A value of 0.5 means 50% occupancy.
    */
-  distinct_hash_join(cudf::table_view const& build,
+  distinct_hash_join(cudf::table_view const& right,
                      cudf::null_equality compare_nulls,
                      double load_factor,
                      rmm::cuda_stream_view stream);
@@ -131,15 +131,15 @@ class distinct_hash_join {
    */
   std::pair<std::unique_ptr<rmm::device_uvector<size_type>>,
             std::unique_ptr<rmm::device_uvector<size_type>>>
-  inner_join(cudf::table_view const& probe,
+  inner_join(cudf::table_view const& left,
              rmm::cuda_stream_view stream,
              rmm::device_async_resource_ref mr) const;
 
   /**
    * @copydoc cudf::distinct_hash_join::left_join
    */
   std::unique_ptr<rmm::device_uvector<size_type>> left_join(
-    cudf::table_view const& probe,
+    cudf::table_view const& left,
     rmm::cuda_stream_view stream,
     rmm::device_async_resource_ref mr) const;
 
@@ -156,11 +156,11 @@ class distinct_hash_join {
                                            rmm::mr::polymorphic_allocator<char>,
                                            cuco_storage_type>;
 
-  bool _has_nested_columns;  ///< True if nested columns are present in build and probe tables
+  bool _has_nested_columns;  ///< True if nested columns are present in right and left tables
   cudf::null_equality _nulls_equal;  ///< Whether to consider nulls as equal
-  cudf::table_view _build;           ///< Input table to build the hash map
+  cudf::table_view _right;           ///< Input table to build the hash map
   std::shared_ptr<cudf::detail::row::equality::preprocessed_table>
-    _preprocessed_build;        ///< Input table preprocssed for row operators
-  hash_table_type _hash_table;  ///< Hash table built on `_build`
+    _preprocessed_right;        ///< Input table preprocssed for row operators
+  hash_table_type _hash_table;  ///< Hash table built on `_right`
 };
 }  // namespace cudf::detail

cpp/include/cudf/detail/join/hash_join.hpp

@@ -29,8 +29,8 @@ class preprocessed_table;
 namespace cudf {
 namespace detail {
 /**
- * @brief Hash join that builds hash table in creation and probes results in subsequent `*_join`
- * member functions.
+ * @brief Hash join that builds a hash table with the right table on construction and probes
+ * results in subsequent `*_join` member functions.
  *
  * User-defined hash function can be passed via the template parameter `Hasher`
  *
@@ -50,17 +50,17 @@ class hash_join {
   hash_join& operator=(hash_join&&)      = delete;
 
   /**
-   * @brief Constructor that internally builds the hash table based on the given `build` table.
+   * @brief Constructor that internally builds the hash table from the given `right` table.
    *
-   * @throw cudf::logic_error if the number of columns in `build` table is 0.
+   * @throw cudf::logic_error if the number of columns in `right` table is 0.
    *
-   * @param build The build table, from which the hash table is built.
-   * @param has_nulls Flag to indicate if the there exists any nulls in the `build` table or
-   *        any `probe` table that will be used later for join.
+   * @param right The right table, from which the hash table is built.
+   * @param has_nulls Flag to indicate if the there exists any nulls in the `right` table or
+   *        any `left` table that will be used later for join.
    * @param compare_nulls Controls whether null join-key values should match or not.
    * @param stream CUDA stream used for device memory operations and kernel launches.
    */
-  hash_join(cudf::table_view const& build,
+  hash_join(cudf::table_view const& right,
             bool has_nulls,
             cudf::null_equality compare_nulls,
             rmm::cuda_stream_view stream);
@@ -70,7 +70,7 @@ class hash_join {
    *
    * @param load_factor The hash table occupancy ratio in (0,1]. A value of 0.5 means 50% occupancy.
    */
-  hash_join(cudf::table_view const& build,
+  hash_join(cudf::table_view const& right,
             bool has_nulls,
             cudf::null_equality compare_nulls,
             double load_factor,
@@ -81,7 +81,7 @@ class hash_join {
    */
   [[nodiscard]] std::pair<std::unique_ptr<rmm::device_uvector<size_type>>,
                           std::unique_ptr<rmm::device_uvector<size_type>>>
-  inner_join(cudf::table_view const& probe,
+  inner_join(cudf::table_view const& left,
              std::optional<std::size_t> output_size,
              rmm::cuda_stream_view stream,
              rmm::device_async_resource_ref mr) const;
@@ -91,7 +91,7 @@ class hash_join {
    */
   [[nodiscard]] std::pair<std::unique_ptr<rmm::device_uvector<size_type>>,
                           std::unique_ptr<rmm::device_uvector<size_type>>>
-  left_join(cudf::table_view const& probe,
+  left_join(cudf::table_view const& left,
             std::optional<std::size_t> output_size,
             rmm::cuda_stream_view stream,
             rmm::device_async_resource_ref mr) const;
@@ -101,83 +101,83 @@ class hash_join {
    */
   [[nodiscard]] std::pair<std::unique_ptr<rmm::device_uvector<size_type>>,
                           std::unique_ptr<rmm::device_uvector<size_type>>>
-  full_join(cudf::table_view const& probe,
+  full_join(cudf::table_view const& left,
             std::optional<std::size_t> output_size,
             rmm::cuda_stream_view stream,
             rmm::device_async_resource_ref mr) const;
 
   /**
    * @copydoc cudf::hash_join::inner_join_size
    */
-  [[nodiscard]] std::size_t inner_join_size(cudf::table_view const& probe,
+  [[nodiscard]] std::size_t inner_join_size(cudf::table_view const& left,
                                             rmm::cuda_stream_view stream) const;
 
   /**
    * @copydoc cudf::hash_join::left_join_size
    */
-  [[nodiscard]] std::size_t left_join_size(cudf::table_view const& probe,
+  [[nodiscard]] std::size_t left_join_size(cudf::table_view const& left,
                                            rmm::cuda_stream_view stream) const;
 
   /**
    * @copydoc cudf::hash_join::full_join_size
    */
-  [[nodiscard]] std::size_t full_join_size(cudf::table_view const& probe,
+  [[nodiscard]] std::size_t full_join_size(cudf::table_view const& left,
                                            rmm::cuda_stream_view stream,
                                            rmm::device_async_resource_ref mr) const;
 
   /**
    * @copydoc cudf::hash_join::inner_join_match_context
    */
   [[nodiscard]] cudf::join_match_context inner_join_match_context(
-    cudf::table_view const& probe,
+    cudf::table_view const& left,
     rmm::cuda_stream_view stream,
     rmm::device_async_resource_ref mr) const;
 
   /**
    * @copydoc cudf::hash_join::left_join_match_context
    */
   [[nodiscard]] cudf::join_match_context left_join_match_context(
-    cudf::table_view const& probe,
+    cudf::table_view const& left,
     rmm::cuda_stream_view stream,
     rmm::device_async_resource_ref mr) const;
 
   /**
    * @copydoc cudf::hash_join::full_join_match_context
    */
   [[nodiscard]] cudf::join_match_context full_join_match_context(
-    cudf::table_view const& probe,
+    cudf::table_view const& left,
     rmm::cuda_stream_view stream,
     rmm::device_async_resource_ref mr) const;
 
  private:
   bool const _is_empty;   ///< true if `_hash_table` is empty
-  bool const _has_nulls;  ///< true if nulls are present in either build table or any probe table
+  bool const _has_nulls;  ///< true if nulls are present in either right table or any left table
   cudf::null_equality const _nulls_equal;  ///< whether to consider nulls as equal
-  cudf::table_view _build;                 ///< input table to build the hash map
+  cudf::table_view _right;                 ///< input table to build the hash map
   std::shared_ptr<cudf::detail::row::equality::preprocessed_table>
-    _preprocessed_build;        ///< input table preprocssed for row operators
+    _preprocessed_right;        ///< input table preprocssed for row operators
   std::unique_ptr<impl> _impl;  ///< CUDA hash table implementation
 
   [[nodiscard]] std::unique_ptr<rmm::device_uvector<size_type>> make_match_counts(
     join_kind join,
-    cudf::table_view const& probe,
+    cudf::table_view const& left,
     rmm::cuda_stream_view stream,
     rmm::device_async_resource_ref mr) const;
 
   template <join_kind Join>
   [[nodiscard]] std::pair<std::unique_ptr<rmm::device_uvector<size_type>>,
                           std::unique_ptr<rmm::device_uvector<size_type>>>
-  join_retrieve(cudf::table_view const& probe,
+  join_retrieve(cudf::table_view const& left,
                 std::optional<std::size_t> output_size,
                 rmm::cuda_stream_view stream,
                 rmm::device_async_resource_ref mr) const;
 
   template <join_kind Join>
-  [[nodiscard]] std::size_t join_size(cudf::table_view const& probe,
+  [[nodiscard]] std::size_t join_size(cudf::table_view const& left,
                                       rmm::cuda_stream_view stream) const;
 
   template <join_kind Join>
-  [[nodiscard]] std::size_t join_size(cudf::table_view const& probe,
+  [[nodiscard]] std::size_t join_size(cudf::table_view const& left,
                                       rmm::cuda_stream_view stream,
                                       rmm::device_async_resource_ref mr) const;
 };