Merge remote-tracking branch 'upstream/dev' into fix-int-namespace

Author: PointKernel

CMakeLists.txt

@@ -88,6 +88,12 @@ target_include_directories(cuco INTERFACE
 target_link_libraries(cuco INTERFACE CCCL::CCCL CUDA::toolkit)
 target_compile_features(cuco INTERFACE cxx_std_17 cuda_std_17)
 
+###################################################################################################
+# - Optionally download RoaringFormatSpec test data -----------------------------------------------
+
+option(CUCO_DOWNLOAD_ROARING_TESTDATA "Download RoaringFormatSpec test data" ON)
+include(${CMAKE_CURRENT_LIST_DIR}/cmake/roaring_testdata.cmake)
+
 ###################################################################################################
 # - optionally build tests ------------------------------------------------------------------------
 

README.md

@@ -28,7 +28,7 @@ function(ConfigureBench BENCH_NAME)
     target_include_directories(${BENCH_NAME} PRIVATE
                                              "${CMAKE_CURRENT_SOURCE_DIR}")
     target_compile_options(${BENCH_NAME} PRIVATE --compiler-options=-Wall --compiler-options=-Wextra
-      --compiler-options=-Werror -Wno-deprecated-gpu-targets --expt-extended-lambda)
+      --compiler-options=-Werror -Wno-deprecated-gpu-targets --expt-extended-lambda -lineinfo)
     # Add GCC-specific warning suppression only for GCC
     if(CMAKE_CXX_COMPILER_ID STREQUAL "GNU")
       target_compile_options(${BENCH_NAME} PRIVATE -Xcompiler -Wno-subobject-linkage)
@@ -104,3 +104,8 @@ ConfigureBench(HYPERLOGLOG_BENCH
 ConfigureBench(BLOOM_FILTER_BENCH
   bloom_filter/add_bench.cu
   bloom_filter/contains_bench.cu)
+
+###################################################################################################
+# - roaring_bitmap benchmarks ---------------------------------------------------------------------
+ConfigureBench(ROARING_BITMAP_BENCH
+  roaring_bitmap/contains_bench.cu)

benchmarks/CMakeLists.txt

@@ -0,0 +1,96 @@
+/*
+ * Copyright (c) 2025, 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
+ *
+ *     http://www.apache.org/licenses/LICENSE-2.0
+ *
+ * Unless required by applicable law or agreed to in writing, software
+ * distributed under the License is distributed on an "AS IS" BASIS,
+ * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied.
+ * See the License for the specific language governing permissions and
+ * limitations under the License.
+ */
+
+#include <benchmark_defaults.hpp>
+#include <benchmark_utils.hpp>
+
+#include <cuco/roaring_bitmap.cuh>
+#include <cuco/utility/key_generator.cuh>
+
+#include <nvbench/nvbench.cuh>
+
+#include <cuda/std/cstddef>
+#include <cuda/std/cstdint>
+#include <thrust/device_vector.h>
+#include <thrust/universal_vector.h>
+
+#include <filesystem>
+#include <fstream>
+#include <string>
+
+using namespace cuco::benchmark;  // defaults
+using namespace cuco::utility;    // key_generator, distribution
+
+template <typename T>
+void roaring_bitmap_contains(nvbench::state& state, nvbench::type_list<T>)
+{
+  auto const num_items   = state.get_int64("NumInputs");
+  auto const bitmap_file = state.get_string_or_default("BitmapFile", {});
+
+  std::ifstream file(bitmap_file, std::ios::binary);
+  if (!file.is_open()) { state.skip("Bitmap file not found"); }
+
+  // Get file size
+  auto const file_size = std::filesystem::file_size(bitmap_file);
+
+  thrust::universal_host_pinned_vector<cuda::std::byte> buffer(file_size);
+
+  file.read(reinterpret_cast<char*>(thrust::raw_pointer_cast(buffer.data())), file_size);
+  file.close();
+
+  cuco::experimental::roaring_bitmap<T> roaring_bitmap(thrust::raw_pointer_cast(buffer.data()));
+
+  thrust::device_vector<T> items(num_items);
+
+  key_generator gen{};
+  gen.generate(distribution::unique{}, items.begin(), items.end());
+
+  thrust::device_vector<bool> contained(items.size(), false);
+
+  state.add_element_count(items.size());
+  state.add_global_memory_reads<T>(items.size(), "InputSize");
+
+  auto& summ = state.add_summary("BitmapSizeMB");
+  summ.set_string("hint", "BitmapSize");
+  summ.set_string("short_name", "BitmapSizeMB");
+  summ.set_string("description", "Bitmap size in MB");
+  summ.set_float64("value", static_cast<double>(file_size) / (1024 * 1024));
+
+  state.exec([&](nvbench::launch& launch) {
+    roaring_bitmap.contains_async(
+      items.begin(), items.end(), contained.begin(), {launch.get_stream()});
+  });
+}
+
+NVBENCH_BENCH_TYPES(roaring_bitmap_contains,
+                    NVBENCH_TYPE_AXES(nvbench::type_list<nvbench::uint32_t>))
+  .set_name("roaring_bitmap_contains")
+  .add_int64_power_of_two_axis("NumInputs", {32})
+// Default benchmark is only available if the Roaring bitmap testdata has been downloaded
+#ifdef CUCO_ROARING_DATA_DIR
+  .add_string_axis("BitmapFile", {std::string(CUCO_ROARING_DATA_DIR) + "/bitmapwithruns.bin"})
+#endif
+  .set_max_noise(cuco::benchmark::defaults::MAX_NOISE);
+
+NVBENCH_BENCH_TYPES(roaring_bitmap_contains,
+                    NVBENCH_TYPE_AXES(nvbench::type_list<nvbench::uint64_t>))
+  .set_name("roaring_bitmap_contains")
+  .add_int64_power_of_two_axis("NumInputs", {31})
+// Default benchmark is only available if the Roaring bitmap testdata has been downloaded
+#ifdef CUCO_ROARING_DATA_DIR
+  .add_string_axis("BitmapFile", {std::string(CUCO_ROARING_DATA_DIR) + "/portable_bitmap64.bin"})
+#endif
+  .set_max_noise(cuco::benchmark::defaults::MAX_NOISE);

benchmarks/roaring_bitmap/contains_bench.cu

@@ -37,6 +37,9 @@ resolve_path() {
 # Ensure the script is being executed in its containing directory
 cd "$( cd "$( dirname "${BASH_SOURCE[0]}" )" && pwd )";
 
+# Determine repo root as the parent of the `ci` directory
+REPO_ROOT="$(cd .. && pwd)"
+
 # Script defaults
 BUILD_TESTS=${BUILD_TESTS:-OFF}
 BUILD_EXAMPLES=${BUILD_EXAMPLES:-OFF}
@@ -51,6 +54,13 @@ HOST_COMPILER=${CXX:-g++} # $CXX if set, otherwise `g++`
 CUDA_ARCHS=native # detect system's GPU architectures
 CXX_STANDARD=17
 
+# Initialize CMAKE_ARGS from environment variable if available
+if [ -n "${CMAKE_ARGS:-}" ]; then
+    read -ra CMAKE_ARGS <<< "$CMAKE_ARGS"
+else
+    CMAKE_ARGS=()
+fi
+
 function usage {
     echo "cuCollections build script"
     echo "Usage: $0 [OPTIONS]"
@@ -103,6 +113,16 @@ function usage {
     echo "    Enables verbose mode for detailed output and builds with C++17 standard."
     echo "    Build files will be written to <repo_root>/build/local and symlinked to <repo_root>/build/latest."
     echo
+    echo "  Using CMAKE_ARGS Environment Variable:"
+    echo "    $ CMAKE_ARGS=\"-DCMAKE_VERBOSE_MAKEFILE=ON -DENABLE_FEATURE=ON\" $0 -t"
+    echo "    $ export CMAKE_ARGS=\"-DCMAKE_VERBOSE_MAKEFILE=ON -DENABLE_FEATURE=ON\""
+    echo "    $ $0 -t"
+    echo "    Uses CMAKE_ARGS environment variable to pass additional CMake options."
+    echo "    Can be overridden by using -- followed by specific arguments."
+    echo
+    echo "  Pass-through to CMake:"
+    echo "    -- [CMake args...]  Anything after -- is forwarded to CMake (overrides CMAKE_ARGS env var)"
+    echo
     exit 1
 }
 
@@ -126,6 +146,7 @@ while [ "${#args[@]}" -ne 0 ]; do
     --arch) CUDA_ARCHS="${args[1]}";    args=("${args[@]:2}");;
     --std)  CXX_STANDARD="${args[1]}";  args=("${args[@]:2}");;
     -v | -verbose | --verbose) VERBOSE=1; args=("${args[@]:1}");;
+    --) CMAKE_ARGS=("${args[@]:1}"); break;;
     -h | -help | --help) usage ;;
     *) echo "Unrecognized option: ${args[0]}"; usage ;;
     esac
@@ -156,14 +177,12 @@ if [ "$BUILD_TESTS" == "OFF" ] && [ "$BUILD_EXAMPLES" == "OFF" ] && [ "$BUILD_BE
     BUILD_BENCHMARKS=ON
 fi
 
+BUILD_DIR="$BUILD_PREFIX/$BUILD_INFIX"
 # Trigger clean (re-)build
 if [ "$CLEAN_BUILD" -eq 1 ]; then
     rm -rf BUILD_DIR
 fi
-
-BUILD_DIR="$BUILD_PREFIX/$BUILD_INFIX"
 mkdir -p $BUILD_DIR
-export BUILD_DIR # TODO remove
 
 # The most recent build will be symlinked to cuCollections/build/latest
 rm -f $BUILD_PREFIX/latest
@@ -186,12 +205,13 @@ CMAKE_OPTIONS="
     -DBUILD_TESTS=${BUILD_TESTS} \
     -DBUILD_EXAMPLES=${BUILD_EXAMPLES} \
     -DBUILD_BENCHMARKS=${BUILD_BENCHMARKS} \
+    ${CMAKE_ARGS[*]}
 "
 
-echo "========================================"
-echo "-- START: $(date)"
-echo "-- GIT_SHA: $(git rev-parse HEAD 2>/dev/null || echo 'Not a repository')"
-echo "-- PWD: $(pwd)"
+echo "[INFO]=============================================="
+echo "-- TIMESTAMP: $(date -u +"%Y-%m-%d %H:%M:%S UTC")"
+echo "-- GIT_SHA: $(git rev-parse HEAD 2>/dev/null || echo 'N/A')"
+echo "-- SRC_DIR: ${REPO_ROOT}"
 echo "-- BUILD_DIR: ${BUILD_DIR}"
 echo "-- BUILD_TYPE: ${BUILD_TYPE}"
 echo "-- PARALLEL_LEVEL: ${PARALLEL_LEVEL}"
@@ -200,21 +220,29 @@ echo "-- BUILD_TESTS: ${BUILD_TESTS}"
 echo "-- BUILD_EXAMPLES: ${BUILD_EXAMPLES}"
 echo "-- BUILD_BENCHMARKS: ${BUILD_BENCHMARKS}"
 
+if [ ${#CMAKE_ARGS[@]} -gt 0 ]; then
+    echo "-- CMAKE_ARGS: ${CMAKE_ARGS[*]}"
+else
+    echo "-- CMAKE_ARGS: (none)"
+fi
+
 # configure
+echo "[CONFIGURE]========================================"
 cmake -S .. -B $BUILD_DIR $CMAKE_OPTIONS
-echo "========================================"
 
 if command -v sccache >/dev/null; then
     source "./sccache_stats.sh" start
+else
+    echo "sccache stats: N/A"
 fi
 
 #build
+echo "[BUILD]============================================"
 cmake --build $BUILD_DIR --parallel $PARALLEL_LEVEL
-echo "========================================"
 echo "Build complete"
 
 if command -v sccache >/dev/null; then
     source "./sccache_stats.sh" end
 else
     echo "sccache stats: N/A"
-fi
+fi

ci/build.sh

@@ -0,0 +1,41 @@
+# =============================================================================
+# Copyright (c) 2025, 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
+#
+# http://www.apache.org/licenses/LICENSE-2.0
+#
+# Unless required by applicable law or agreed to in writing, software distributed under the License
+# is distributed on an "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express
+# or implied. See the License for the specific language governing permissions and limitations under
+# the License.
+# =============================================================================
+
+# Only act if enabled
+if(NOT CUCO_DOWNLOAD_ROARING_TESTDATA)
+  return()
+endif()
+
+set(CUCO_ROARING_DATA_DIR "${CMAKE_BINARY_DIR}/data/roaring_bitmap")
+
+file(MAKE_DIRECTORY "${CUCO_ROARING_DATA_DIR}")
+
+set(ROARING_FORMATSPEC_BASE "https://raw.githubusercontent.com/RoaringBitmap/RoaringFormatSpec/5177ad9")
+
+rapids_cmake_download_with_retry("${ROARING_FORMATSPEC_BASE}/testdata/bitmapwithoutruns.bin"
+                                 "${CUCO_ROARING_DATA_DIR}/bitmapwithoutruns.bin"
+                                 "d719ae2e0150a362ef7cf51c361527585891f01460b1a92bcfb6a7257282a442")
+
+rapids_cmake_download_with_retry("${ROARING_FORMATSPEC_BASE}/testdata/bitmapwithruns.bin"
+                                 "${CUCO_ROARING_DATA_DIR}/bitmapwithruns.bin"
+                                 "1f1909bfdd354fa2f0694fe88b8076833ca5383ad9fc3f68f2709c84a2ab70e3")
+
+rapids_cmake_download_with_retry("${ROARING_FORMATSPEC_BASE}/testdata64/portable_bitmap64.bin"
+                                 "${CUCO_ROARING_DATA_DIR}/portable_bitmap64.bin"
+                                 "b5a553a759167f5f9ccb3fa21552d943b4c73235635b753376f4faf62067d178")
+
+message(STATUS "Roaring Bitmap test data downloaded to: ${CUCO_ROARING_DATA_DIR}")
+
+# Define macro only when data is available
+add_compile_definitions(CUCO_ROARING_DATA_DIR="${CUCO_ROARING_DATA_DIR}")

cmake/roaring_testdata.cmake

@@ -52,3 +52,4 @@ ConfigureExample(STATIC_MULTIMAP_HOST_BULK_EXAMPLE "${CMAKE_CURRENT_SOURCE_DIR}/
 ConfigureExample(HYPERLOGLOG_HOST_BULK_EXAMPLE "${CMAKE_CURRENT_SOURCE_DIR}/hyperloglog/host_bulk_example.cu")
 ConfigureExample(HYPERLOGLOG_DEVICE_REF_EXAMPLE "${CMAKE_CURRENT_SOURCE_DIR}/hyperloglog/device_ref_example.cu")
 ConfigureExample(BLOOM_FILTER_HOST_BULK_EXAMPLE "${CMAKE_CURRENT_SOURCE_DIR}/bloom_filter/host_bulk_example.cu")
+ConfigureExample(ROARING_BITMAP_HOST_BULK_EXAMPLE "${CMAKE_CURRENT_SOURCE_DIR}/roaring_bitmap/host_bulk_example.cu")