Migrate hashing operations to pylibcudf

conda/environments/all_cuda-118_arch-x86_64.yaml

@@ -46,6 +46,7 @@ dependencies:
 - librdkafka>=2.5.0,<2.6.0a0
 - librmm==24.12.*,>=0.0.0a0
 - make
+- mmh3
 - moto>=4.0.8
 - msgpack-python
 - myst-nb
@@ -76,6 +77,7 @@ dependencies:
 - pytest-xdist
 - pytest<8
 - python-confluent-kafka>=2.5.0,<2.6.0a0
+- python-xxhash
 - python>=3.10,<3.13
 - pytorch>=2.1.0
 - rapids-build-backend>=0.3.0,<0.4.0.dev0

conda/environments/all_cuda-125_arch-x86_64.yaml

@@ -45,6 +45,7 @@ dependencies:
 - librdkafka>=2.5.0,<2.6.0a0
 - librmm==24.12.*,>=0.0.0a0
 - make
+- mmh3
 - moto>=4.0.8
 - msgpack-python
 - myst-nb
@@ -74,6 +75,7 @@ dependencies:
 - pytest-xdist
 - pytest<8
 - python-confluent-kafka>=2.5.0,<2.6.0a0
+- python-xxhash
 - python>=3.10,<3.13
 - pytorch>=2.1.0
 - rapids-build-backend>=0.3.0,<0.4.0.dev0

cpp/include/cudf/hashing.hpp

@@ -22,12 +22,6 @@
 
 namespace CUDF_EXPORT cudf {
 
-/**
- * @addtogroup column_hash
- * @{
- * @file
- */
-
 /**
  * @brief Type of hash value
  *
@@ -42,6 +36,12 @@ static constexpr uint32_t DEFAULT_HASH_SEED = 0;
 //! Hash APIs
 namespace hashing {
 
+/**
+ * @addtogroup column_hash
+ * @{
+ * @file
+ */
+
 /**
  * @brief Computes the MurmurHash3 32-bit hash value of each row in the given table
  *
@@ -183,7 +183,8 @@ std::unique_ptr<column> xxhash_64(
   rmm::cuda_stream_view stream      = cudf::get_default_stream(),
   rmm::device_async_resource_ref mr = cudf::get_current_device_resource_ref());
 
+/** @} */  // end of group
+
 }  // namespace hashing
 
-/** @} */  // end of group
 }  // namespace CUDF_EXPORT cudf

cpp/src/hash/md5_hash.cu

@@ -302,7 +302,8 @@ std::unique_ptr<column> md5(table_view const& input,
                              }
                              return md5_leaf_type_check(col.type());
                            }),
-               "Unsupported column type for hash function.");
+               "Unsupported column type for hash function.",
+               cudf::data_type_error);
 
   // Digest size in bytes
   auto constexpr digest_size = 32;

cpp/src/hash/sha_hash.cuh

@@ -513,7 +513,8 @@ std::unique_ptr<column> sha_hash(table_view const& input,
   CUDF_EXPECTS(
     std::all_of(
       input.begin(), input.end(), [](auto const& col) { return sha_leaf_type_check(col.type()); }),
-    "Unsupported column type for hash function.");
+    "Unsupported column type for hash function.",
+    cudf::data_type_error);
 
   // Result column allocation and creation
   auto begin = thrust::make_constant_iterator(Hasher::digest_size);

cpp/tests/hashing/sha1_test.cpp

@@ -137,7 +137,7 @@ TEST_F(SHA1HashTest, ListsUnsupported)
 
   auto const input = cudf::table_view({strings_list_col});
 
-  EXPECT_THROW(cudf::hashing::sha1(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha1(input), cudf::data_type_error);
 }
 
 TEST_F(SHA1HashTest, StructsUnsupported)
@@ -146,7 +146,7 @@ TEST_F(SHA1HashTest, StructsUnsupported)
   auto struct_col  = cudf::test::structs_column_wrapper{{child_col}};
   auto const input = cudf::table_view({struct_col});
 
-  EXPECT_THROW(cudf::hashing::sha1(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha1(input), cudf::data_type_error);
 }
 
 template <typename T>

cpp/tests/hashing/sha224_test.cpp

@@ -137,7 +137,7 @@ TEST_F(SHA224HashTest, ListsUnsupported)
 
   auto const input = cudf::table_view({strings_list_col});
 
-  EXPECT_THROW(cudf::hashing::sha224(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha224(input), cudf::data_type_error);
 }
 
 TEST_F(SHA224HashTest, StructsUnsupported)
@@ -146,7 +146,7 @@ TEST_F(SHA224HashTest, StructsUnsupported)
   auto struct_col  = cudf::test::structs_column_wrapper{{child_col}};
   auto const input = cudf::table_view({struct_col});
 
-  EXPECT_THROW(cudf::hashing::sha224(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha224(input), cudf::data_type_error);
 }
 
 template <typename T>

cpp/tests/hashing/sha256_test.cpp

@@ -136,7 +136,7 @@ TEST_F(SHA256HashTest, ListsUnsupported)
 
   auto const input = cudf::table_view({strings_list_col});
 
-  EXPECT_THROW(cudf::hashing::sha256(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha256(input), cudf::data_type_error);
 }
 
 TEST_F(SHA256HashTest, StructsUnsupported)
@@ -145,7 +145,7 @@ TEST_F(SHA256HashTest, StructsUnsupported)
   auto struct_col  = cudf::test::structs_column_wrapper{{child_col}};
   auto const input = cudf::table_view({struct_col});
 
-  EXPECT_THROW(cudf::hashing::sha256(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha256(input), cudf::data_type_error);
 }
 
 template <typename T>

cpp/tests/hashing/sha384_test.cpp

@@ -155,7 +155,7 @@ TEST_F(SHA384HashTest, ListsUnsupported)
 
   auto const input = cudf::table_view({strings_list_col});
 
-  EXPECT_THROW(cudf::hashing::sha384(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha384(input), cudf::data_type_error);
 }
 
 TEST_F(SHA384HashTest, StructsUnsupported)
@@ -164,7 +164,7 @@ TEST_F(SHA384HashTest, StructsUnsupported)
   auto struct_col  = cudf::test::structs_column_wrapper{{child_col}};
   auto const input = cudf::table_view({struct_col});
 
-  EXPECT_THROW(cudf::hashing::sha384(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha384(input), cudf::data_type_error);
 }
 
 template <typename T>

cpp/tests/hashing/sha512_test.cpp

@@ -155,7 +155,7 @@ TEST_F(SHA512HashTest, ListsUnsupported)
 
   auto const input = cudf::table_view({strings_list_col});
 
-  EXPECT_THROW(cudf::hashing::sha512(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha512(input), cudf::data_type_error);
 }
 
 TEST_F(SHA512HashTest, StructsUnsupported)
@@ -164,7 +164,7 @@ TEST_F(SHA512HashTest, StructsUnsupported)
   auto struct_col  = cudf::test::structs_column_wrapper{{child_col}};
   auto const input = cudf::table_view({struct_col});
 
-  EXPECT_THROW(cudf::hashing::sha512(input), cudf::logic_error);
+  EXPECT_THROW(cudf::hashing::sha512(input), cudf::data_type_error);
 }
 
 template <typename T>

dependencies.yaml

@@ -828,6 +828,7 @@ dependencies:
           - pytest-benchmark
           - pytest-cases>=3.8.2
           - scipy
+          - mmh3
       - output_types: conda
         packages:
           - aiobotocore>=2.2.0
@@ -836,12 +837,14 @@ dependencies:
           - msgpack-python
           - moto>=4.0.8
           - s3fs>=2022.3.0
-      - output_types: pyproject
+          - python-xxhash
+      - output_types: [pyproject, requirements]
         packages:
           - msgpack
           - &tokenizers tokenizers==0.15.2
           - &transformers transformers==4.39.3
           - tzdata
+          - xxhash
     specific:
       - output_types: [conda, requirements]
         matrices:

docs/cudf/source/user_guide/api_docs/pylibcudf/hashing.rst

@@ -0,0 +1,6 @@
+=======
+hashing
+=======
+
+.. automodule:: pylibcudf.hashing
+    :members:

docs/cudf/source/user_guide/api_docs/pylibcudf/index.rst

@@ -19,6 +19,7 @@ This page provides API documentation for pylibcudf.
     filling
     gpumemoryview
     groupby
+    hashing
     interop
     join
     json

python/cudf/cudf/_lib/hash.pyx

@@ -1,27 +1,12 @@
 # Copyright (c) 2020-2024, NVIDIA CORPORATION.
 
-from cudf.core.buffer import acquire_spill_lock
+import pylibcudf as plc
 
-from libcpp.memory cimport unique_ptr
-from libcpp.utility cimport move
+from cudf.core.buffer import acquire_spill_lock
 
-from pylibcudf.libcudf.column.column cimport column
-from pylibcudf.libcudf.hash cimport (
-    md5,
-    murmurhash3_x86_32,
-    sha1,
-    sha224,
-    sha256,
-    sha384,
-    sha512,
-    xxhash_64,
-)
-from pylibcudf.libcudf.table.table_view cimport table_view
+from pylibcudf.table cimport Table
 
 from cudf._lib.column cimport Column
-from cudf._lib.utils cimport table_view_from_columns
-
-import pylibcudf as plc
 
 
 @acquire_spill_lock()
@@ -37,32 +22,26 @@ def hash_partition(list source_columns, list columns_to_hash,
 
 @acquire_spill_lock()
 def hash(list source_columns, str method, int seed=0):
-    cdef table_view c_source_view = table_view_from_columns(source_columns)
-    cdef unique_ptr[column] c_result
+    cdef Table ctbl = Table(
+        [c.to_pylibcudf(mode="read") for c in source_columns]
+    )
     if method == "murmur3":
-        with nogil:
-            c_result = move(murmurhash3_x86_32(c_source_view, seed))
+        return Column.from_pylibcudf(plc.hashing.murmurhash3_x86_32(ctbl, seed))
+    elif method == "xxhash64":
+        return Column.from_pylibcudf(plc.hashing.xxhash_64(ctbl, seed))
     elif method == "md5":
-        with nogil:
-            c_result = move(md5(c_source_view))
+        return Column.from_pylibcudf(plc.hashing.md5(ctbl))
     elif method == "sha1":
-        with nogil:
-            c_result = move(sha1(c_source_view))
+        return Column.from_pylibcudf(plc.hashing.sha1(ctbl))
     elif method == "sha224":
-        with nogil:
-            c_result = move(sha224(c_source_view))
+        return Column.from_pylibcudf(plc.hashing.sha224(ctbl))
     elif method == "sha256":
-        with nogil:
-            c_result = move(sha256(c_source_view))
+        return Column.from_pylibcudf(plc.hashing.sha256(ctbl))
     elif method == "sha384":
-        with nogil:
-            c_result = move(sha384(c_source_view))
+        return Column.from_pylibcudf(plc.hashing.sha384(ctbl))
     elif method == "sha512":
-        with nogil:
-            c_result = move(sha512(c_source_view))
-    elif method == "xxhash64":
-        with nogil:
-            c_result = move(xxhash_64(c_source_view, seed))
+        return Column.from_pylibcudf(plc.hashing.sha512(ctbl))
     else:
-        raise ValueError(f"Unsupported hash function: {method}")
-    return Column.from_unique_ptr(move(c_result))
+        raise ValueError(
+            f"Unsupported hashing algorithm {method}."
+        )

python/cudf/pyproject.toml

@@ -53,6 +53,7 @@ test = [
     "cramjam",
     "fastavro>=0.22.9",
     "hypothesis",
+    "mmh3",
     "msgpack",
     "pytest-benchmark",
     "pytest-cases>=3.8.2",
@@ -63,6 +64,7 @@ test = [
     "tokenizers==0.15.2",
     "transformers==4.39.3",
     "tzdata",
+    "xxhash",
 ] # This list was generated by `rapids-dependency-file-generator`. To make changes, edit ../../dependencies.yaml and run `rapids-dependency-file-generator`.
 pandas-tests = [
     "ipython",

python/pylibcudf/pylibcudf/CMakeLists.txt

@@ -26,6 +26,7 @@ set(cython_sources
     filling.pyx
     gpumemoryview.pyx
     groupby.pyx
+    hashing.pyx
     interop.pyx
     join.pyx
     json.pyx

python/pylibcudf/pylibcudf/__init__.pxd

@@ -13,6 +13,7 @@ from . cimport (
     expressions,
     filling,
     groupby,
+    hashing,
     interop,
     join,
     json,
@@ -63,6 +64,7 @@ __all__ = [
     "filling",
     "gpumemoryview",
     "groupby",
+    "hashing",
     "interop",
     "join",
     "json",

python/pylibcudf/pylibcudf/__init__.py

@@ -22,6 +22,7 @@
     expressions,
     filling,
     groupby,
+    hashing,
     interop,
     io,
     join,
@@ -73,6 +74,7 @@
     "filling",
     "gpumemoryview",
     "groupby",
+    "hashing",
     "interop",
     "io",
     "join",

python/pylibcudf/pylibcudf/hashing.pxd

@@ -0,0 +1,30 @@
+# Copyright (c) 2024, NVIDIA CORPORATION.
+
+from libc.stdint cimport uint32_t, uint64_t
+
+from .column cimport Column
+from .table cimport Table
+
+
+cpdef Column murmurhash3_x86_32(
+    Table input,
+    uint32_t seed=*
+)
+
+cpdef Table murmurhash3_x64_128(
+    Table input,
+    uint64_t seed=*
+)
+
+
+cpdef Column xxhash_64(
+    Table input,
+    uint64_t seed=*
+)
+
+cpdef Column md5(Table input)
+cpdef Column sha1(Table input)
+cpdef Column sha224(Table input)
+cpdef Column sha256(Table input)
+cpdef Column sha384(Table input)
+cpdef Column sha512(Table input)