Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Refactor cudf::contains #10997

Merged
merged 346 commits into from
May 31, 2022
Merged
Show file tree
Hide file tree
Changes from 250 commits
Commits
Show all changes
346 commits
Select commit Hold shift + click to select a range
bcd6962
Merge branch 'branch-22.04' into list-row-eq
devavret Mar 23, 2022
0c12c15
Move equality comparator to experimental header
devavret Mar 23, 2022
b41b3fa
Style fixes
devavret Mar 23, 2022
4919b04
Merge branch 'branch-22.06' into list-row-eq
devavret Mar 23, 2022
3dfc133
Review changes
devavret Mar 24, 2022
9031900
Review changes requested by @hyperbolic2346
devavret Mar 24, 2022
119d830
Add an equality comparable check similar to lex comparable check
devavret Mar 24, 2022
1cefa5a
Move linked column to a common header in utilities
devavret Mar 24, 2022
3781d01
Merge branch 'list-row-eq' into list-row-hash
devavret Mar 24, 2022
34aa66b
Review changes
devavret Mar 28, 2022
e7ea7f9
Change to progressive slicing logic
devavret Mar 28, 2022
e60fbd4
pull type dispatcher out of element compare loop
devavret Mar 28, 2022
42319ad
Move slicing logic to lists_column_device_view and new structs_column…
devavret Mar 28, 2022
67c035c
Move list size iterator and make it only constructible from list_colu…
devavret Mar 28, 2022
e5fe24c
push slice logic into columns_device_view
devavret Mar 29, 2022
85861a9
Add validity safe iterator
devavret Mar 29, 2022
9db7479
Move element_range_comparator to element_comparator's private
devavret Mar 29, 2022
db3b79b
style fixes
devavret Mar 29, 2022
6ca7deb
Docs for the newly added stuff
devavret Mar 29, 2022
2fc3d3d
Merge branch 'branch-22.06' into list-row-eq
devavret Mar 29, 2022
1c3a99d
review changes
devavret Mar 29, 2022
a0e581c
review changes
devavret Mar 30, 2022
046c407
Review changes
devavret Mar 30, 2022
6a282a6
linked_column_view inherit from column_view_base
devavret Mar 30, 2022
0f768ac
spell check
devavret Mar 30, 2022
6c64915
Merge branch 'list-row-eq' into list-row-hash
devavret Mar 30, 2022
92c1ff5
Change composition to private inheritance
devavret Mar 31, 2022
4c0e7fa
Replace __host__ __device__ with macro
devavret Mar 31, 2022
75104bb
Add more null frequencies to benchmark
devavret Mar 31, 2022
1e1053b
Templatize make_validity_iterator
devavret Mar 31, 2022
bcfe91b
Increase testing for null frequency
devavret Mar 31, 2022
981438d
curr_col -> temp_col
devavret Mar 31, 2022
5bbf18e
element_range_comparator -> column_comparator
devavret Mar 31, 2022
8e18d66
cleaner column_view conversion
devavret Mar 31, 2022
7207b0b
Merge branch 'list-row-eq' into list-row-hash
devavret Mar 31, 2022
1c525d8
Make distinct work again after merges
devavret Apr 1, 2022
75eaed4
delete copy ctor and assignment operator
devavret Apr 1, 2022
be98357
iterator docs
devavret Apr 1, 2022
9b7f3a0
use owner API for hash
devavret Apr 4, 2022
dda3c1d
Enable nulls not equal path
devavret Apr 4, 2022
370d3b3
Add struct support to hashing
devavret Apr 4, 2022
f4c509a
Handle empty struct in list equality
devavret Apr 8, 2022
d1386cf
Handle empty list (without offsets)
devavret Apr 8, 2022
6aef29f
Merge branch 'branch-22.06' into list-row-eq
devavret Apr 8, 2022
1ca9bcd
Merge branch 'list-row-eq' into list-row-hash
devavret Apr 8, 2022
2c53501
Add seed support to element hasher
devavret Apr 9, 2022
2a988e4
Change murmur hash table to use new row operators
devavret Apr 10, 2022
124c7df
switch over serial hash to experimental::row::hash::element_hasher to…
devavret Apr 10, 2022
7863017
limit changes to distinct and not other stream compaction ops
devavret Apr 10, 2022
8bb7572
small ctor changes in row hasher
devavret Apr 10, 2022
3fb6865
Use accumulate wherever possible
devavret Apr 11, 2022
3cc1159
Merge branch 'branch-22.06' into list-row-eq
devavret Apr 11, 2022
e76a2f3
Add nulls in benchmark
devavret Apr 12, 2022
4f46db9
Add seed support to row hasher
devavret Apr 12, 2022
cbe757a
adding noexcept speeds up a bit
devavret Apr 12, 2022
92c77a5
hide device_row_hasher ctor to disallow use without owning row_hasher
devavret Apr 12, 2022
655bedd
Move nested traversal into an adapter class
devavret Apr 12, 2022
8078e3c
Column_device_view review changes
devavret Apr 12, 2022
aa5385b
Merge branch 'list-row-eq' into list-row-hash
devavret Apr 12, 2022
9c1b0d0
Add empty nesting tests
devavret Apr 12, 2022
e314601
Merge branch 'branch-22.06' into list-row-hash
devavret Apr 13, 2022
b5ca2e7
style check copyright
devavret Apr 13, 2022
0634092
Add a test
ttnghia Apr 14, 2022
7cb22f0
Fix bug that ignores null at the top level
ttnghia Apr 14, 2022
52ef585
Rename variable
ttnghia Apr 14, 2022
1534bc0
Rename variable
ttnghia Apr 14, 2022
2c40182
Review comments by @bdice
devavret Apr 14, 2022
cecf5ec
Add a test
ttnghia Apr 14, 2022
d9d1818
Fix test
ttnghia Apr 14, 2022
48938c5
Doc updates
devavret Apr 14, 2022
bede9db
Review changes
devavret Apr 14, 2022
405975c
Fix empty struct bug
devavret Apr 14, 2022
a7f6963
more review changes
devavret Apr 14, 2022
4388ab4
iterate over structs directly instead of transforming from indices
devavret Apr 14, 2022
0150c47
remove null hash as a param from adapter
devavret Apr 14, 2022
94bb3c1
Implement `table_comparator` in the new experimental struct comparator
ttnghia Apr 14, 2022
1cd15ff
Temporary disable failed tests
ttnghia Apr 14, 2022
8e3689f
Fix slicing problem
ttnghia Apr 15, 2022
65a5cde
review changes
devavret Apr 15, 2022
51789d2
has_nulls -> check_nulls
devavret Apr 15, 2022
023ec1f
Fix a problem with struct offsets being already applied
devavret Apr 15, 2022
1b0238f
leftover has_nulls -> check_nulls, and docs
devavret Apr 17, 2022
4a97709
Change how seed is included in row hashin
devavret Apr 17, 2022
8bdcd9e
change adapter to stored element hasher
devavret Apr 18, 2022
70086b6
Fix broken pytest
devavret Apr 18, 2022
a8a2788
Fix index
ttnghia Apr 18, 2022
c532d7f
Revert "Fix index"
ttnghia Apr 19, 2022
5faa24f
Shorten function
ttnghia Apr 21, 2022
bc15949
Rewrite `search_bound`
ttnghia Apr 22, 2022
53bb891
Add `lexicographic::table_comparator`
ttnghia Apr 22, 2022
652e90d
Rewrite `contains_scalar_dispatch`
ttnghia Apr 22, 2022
0205e4e
Implement scalar search for list type
ttnghia Apr 22, 2022
4c99c81
Add const, and comments
ttnghia Apr 25, 2022
8a3adfa
Merge remote-tracking branch 'devavret/list-row-hash' into support_st…
ttnghia Apr 25, 2022
7b3de03
Add template keyword
ttnghia Apr 26, 2022
b256026
Use `cuco::static_map`
ttnghia Apr 28, 2022
6c9e0e4
Rename template parameter
ttnghia Apr 28, 2022
329ea9b
Use `make_device_uvector` instead of `thrust::uninitialized_fill`
ttnghia Apr 28, 2022
4904a65
Misc
ttnghia Apr 28, 2022
50b8891
Add strong index type.
bdice Apr 16, 2022
b9ed4d7
Revert changes to non-experimental row operators.
bdice Apr 20, 2022
d67f17e
Use enum for strongly typed index.
bdice May 3, 2022
464ed2b
Add two table comparator and adapter.
bdice May 3, 2022
b26b318
Add friends. :)
bdice May 3, 2022
1fd199d
Apply two-table comparator to search algorithms.
bdice May 3, 2022
18bd9f0
Move shared lhs/rhs logic into launch_search.
bdice May 3, 2022
b5b8b39
Improve comments, remove old code.
bdice May 3, 2022
96a72fa
Merge branch 'branch-22.06' into support_structs_in_contains
ttnghia May 4, 2022
6877ad7
Add template list tests
ttnghia May 4, 2022
075d9d6
Merge branch 'branch-22.06' into support_structs_in_contains
ttnghia May 11, 2022
388c516
Merge branch 'strong-index-type' into support_structs_in_contains
ttnghia May 11, 2022
f3dd8e7
Hack to get thing working
ttnghia May 11, 2022
4060b4f
Merge remote-tracking branch 'upstream/branch-22.06' into strong-inde…
bdice May 11, 2022
73c4b27
Move strong typing code into cudf::experimental::row::lexicographic.
bdice May 11, 2022
02774fb
Merge branch 'branch-22.06' into support_structs_in_contains
ttnghia May 13, 2022
f746182
Fix compiler warning treated as error
ttnghia May 13, 2022
ef40471
Fix error
ttnghia May 13, 2022
3a8f5ca
Change headers order
ttnghia May 13, 2022
f1c7079
Merge branch 'strong-index-type' into support_structs_in_contains
ttnghia May 13, 2022
ce81da4
Fix row comparator
ttnghia May 13, 2022
b339138
Adopt new row comparator
ttnghia May 13, 2022
9cdbe27
Merge remote-tracking branch 'upstream/branch-22.06' into strong-inde…
bdice May 13, 2022
c8a38fe
Improve comment.
bdice May 13, 2022
8b5ef34
Fix docstrings.
bdice May 13, 2022
4c63f4c
Fix tests
ttnghia May 13, 2022
1845e5e
Fix comparator
ttnghia May 13, 2022
205c0f2
Cleanup `search_ordered.cu`
ttnghia May 13, 2022
f475cf0
Add `contains_nested.cu` and remove struct `contains.hpp`
ttnghia May 13, 2022
48de04d
Remove `contains.cu` for struct
ttnghia May 13, 2022
0d8a208
Move file
ttnghia May 13, 2022
77f85b4
Enable weak ordering machinery (weak_ordering_comparator_impl) to wra…
bdice May 13, 2022
529e944
Remove template template parameters.
bdice May 13, 2022
fb0e192
Use references.
bdice May 13, 2022
d157199
Remove file
ttnghia May 13, 2022
56d99ba
Use Ts const...
bdice May 13, 2022
c5998b7
Move strong typing to cudf::experimental::row.
bdice May 13, 2022
b78d978
Use constexpr.
bdice May 13, 2022
427ce6e
Rewrite code
ttnghia May 13, 2022
f0f1efc
Add const
ttnghia May 13, 2022
81a4572
Fix compile error
ttnghia May 13, 2022
8b1f4e6
Add type check
ttnghia May 13, 2022
3aea8d4
Use custom iterator class.
bdice May 14, 2022
bbaf360
Use __device__ only.
bdice May 14, 2022
4a1d7aa
Add comment.
bdice May 14, 2022
09c5661
Use symmetry of comparator (now possible with weak ordering) to avoid…
bdice May 14, 2022
290323f
Add constexpr to two_table_device_row_comparator_adapter.
bdice May 14, 2022
4c69edd
Remove forward (always accepts lvalues).
bdice May 14, 2022
ea8c223
Merge branch 'strong-index-type' into strong_typed_index
ttnghia May 16, 2022
857f570
Implement strong typed index for equality comparator
ttnghia May 16, 2022
d9f63f0
Adopt new strong typed index
ttnghia May 16, 2022
12f7a8b
Remove header
ttnghia May 16, 2022
fbd5b90
Indicate reversed signature.
bdice May 16, 2022
3db6484
Move constructor to implementation, add shape compatibility check.
bdice May 16, 2022
3e81b53
Improve docstrings.
bdice May 16, 2022
1834095
Use thrust::iterator_facade.
bdice May 16, 2022
9cb656b
Merge branch 'branch-22.06' into strong_typed_index
ttnghia May 16, 2022
c766bf3
Merge branch 'strong-index-type' into strong_typed_index
ttnghia May 16, 2022
a311bcc
Add type check
ttnghia May 16, 2022
b935835
Change parameter side for comparator
ttnghia May 16, 2022
73b8caf
Add comment
ttnghia May 17, 2022
ff26024
Use const for struct members.
bdice May 17, 2022
f779bff
Slim down the strong index layer by using a templated struct.
bdice May 17, 2022
157abbc
Simplify construction.
bdice May 17, 2022
a2ac19d
Use size_type const where possible.
bdice May 17, 2022
75249e8
Require weakly or strongly typed values for lhs_index and rhs_index.
bdice May 17, 2022
f50faf5
Merge branch 'strong-index-type' into strong_typed_index
ttnghia May 17, 2022
bed1162
Unconstrain template typenames.
bdice May 18, 2022
6930952
Merge branch 'strong-index-type' into strong_typed_index
ttnghia May 18, 2022
2781af1
Remove type check
ttnghia May 18, 2022
8b239d4
Update adapter comparator
ttnghia May 18, 2022
17bd96c
Merge branch 'branch-22.06' into strong_typed_index
ttnghia May 18, 2022
ae77f68
Merge branch 'branch-22.06' into strong_typed_index
ttnghia May 18, 2022
d6b5eb9
Remove deprecated code
ttnghia May 18, 2022
5f39a28
Fix comments
ttnghia May 18, 2022
bf3555c
Fix header check
ttnghia May 18, 2022
df98698
Fix format
ttnghia May 18, 2022
b67d070
Rename variables and fix header
ttnghia May 18, 2022
4429cc4
Merge branch 'branch-22.06' into strong_typed_index
ttnghia May 18, 2022
893db8a
Address review comments
ttnghia May 18, 2022
7e69d3b
Merge branch 'branch-22.06' into strong_typed_index
ttnghia May 18, 2022
c51b053
Fix renaming issue
ttnghia May 18, 2022
934ee73
Switch operands
ttnghia May 18, 2022
cf996f0
Update cpp/src/structs/search/contains.cu
ttnghia May 18, 2022
1dd30eb
Merge branch 'strong_typed_index' into support_structs_in_contains
ttnghia May 18, 2022
73fbc47
Fix compile issue
ttnghia May 18, 2022
fdcbd43
Rewrite `multi_contains_nested_elements`
ttnghia May 18, 2022
261e98c
Add `nested_type_scalar_to_column_view`
ttnghia May 19, 2022
91447cf
Rename .cuh into .hpp
ttnghia May 19, 2022
16ea31d
WIP
ttnghia May 19, 2022
f0f2c4a
Merge remote-tracking branch 'nghia/column_utility' into support_stru…
ttnghia May 19, 2022
662933d
Adopt `nested_type_scalar_to_column_view`
ttnghia May 19, 2022
658cefb
Rewrite `contains_nested.cu`
ttnghia May 19, 2022
29613bc
Move implementation to cpp file
ttnghia May 19, 2022
a596d40
Merge branch 'column_utility' into support_structs_in_contains
ttnghia May 19, 2022
e765cc8
Add comment
ttnghia May 19, 2022
ca25394
Move file
ttnghia May 19, 2022
1b6c3e7
Fix comment
ttnghia May 19, 2022
85e426a
Merge branch 'branch-22.06' into support_structs_in_contains
ttnghia May 20, 2022
5a1d7b3
Add declarations into header
ttnghia May 20, 2022
7fd9b8f
Remove extern declaration
ttnghia May 20, 2022
0246b5a
Add doxygen
ttnghia May 20, 2022
5a3fefe
Rewrite comments
ttnghia May 20, 2022
2c1d8ad
Update doxygen
ttnghia May 20, 2022
1ba60a2
Cleanup
ttnghia May 20, 2022
d72da6e
Simplify code
ttnghia May 20, 2022
5079cce
Add empty test
ttnghia May 20, 2022
427a36d
Add TrivialInputTest
ttnghia May 20, 2022
e1edeb4
Implement `row_hasher_adapter`
ttnghia May 20, 2022
9811c4d
Fix test
ttnghia May 20, 2022
df13c3c
Rewrite comments
ttnghia May 20, 2022
bbbf223
Rewrite comments
ttnghia May 20, 2022
3e94ab6
Add SlicedInputNoNulls test
ttnghia May 20, 2022
5816703
Add SlicedInputHavingNulls test
ttnghia May 20, 2022
d0e2cbf
Add StructOfLists test
ttnghia May 20, 2022
705087b
Cleanup
ttnghia May 20, 2022
008520a
Rename variables
ttnghia May 20, 2022
bae8587
Add TrivialInputTest
ttnghia May 20, 2022
e8c72e6
Add EmptyInputTest
ttnghia May 20, 2022
eff240b
Rename variable
ttnghia May 20, 2022
8f3b640
Add more test
ttnghia May 20, 2022
bde4f2e
Rename tests and change macros
ttnghia May 20, 2022
34afa0f
Remove variable
ttnghia May 20, 2022
c9b3e56
Add more list tests
ttnghia May 20, 2022
07cf7e3
Modify test
ttnghia May 20, 2022
fb87126
Add verbosity
ttnghia May 20, 2022
40754e0
Add TrivialInput test
ttnghia May 20, 2022
cdbaee3
Add SlicedInputHavingNulls test
ttnghia May 20, 2022
c153d38
Add ListsOfStructs test
ttnghia May 20, 2022
f38b1ff
Fix cmake format
ttnghia May 20, 2022
2554afc
Update copyright year
ttnghia May 20, 2022
679524d
Materialize column from scalar
ttnghia May 23, 2022
3678871
Merge branch 'branch-22.08' into support_structs_in_contains
ttnghia May 23, 2022
fa0b92e
Merge branch 'branch-22.08' into support_structs_in_contains
ttnghia May 26, 2022
e57cfd0
Remove `search_list_test.cpp`
ttnghia May 27, 2022
76a8e82
Reverse `search_struct_test.cpp`
ttnghia May 27, 2022
d43babd
Reverse `search.hpp`
ttnghia May 27, 2022
dc2affc
Roll back support for list
ttnghia May 27, 2022
11e0ee9
Remove `multi_contains_nested_elements`
ttnghia May 27, 2022
545ed63
Add declaration for `contains_nested_element`
ttnghia May 27, 2022
ff91ca8
fix compile errors
ttnghia May 27, 2022
b8195d2
Merge branch 'branch-22.08' into refactor_cudfcontains
ttnghia May 27, 2022
4e00bb9
Remove `struct/search/contains.cu`
ttnghia May 27, 2022
e71fea7
Rewrite `contains_nested.cu`
ttnghia May 27, 2022
971cf77
Cleanup headers
ttnghia May 27, 2022
e9c8a63
Simplify code
ttnghia May 27, 2022
059811a
Remove `static_cast`
ttnghia May 27, 2022
436f394
Merge branch 'branch-22.08' into refactor_cudfcontains
ttnghia May 28, 2022
64769c3
Fix missing copydoc
ttnghia May 31, 2022
70e96cf
Merge branch 'branch-22.08' into refactor_cudfcontains
ttnghia May 31, 2022
b676a31
Update cpp/src/search/contains.cu
ttnghia May 31, 2022
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
1 change: 0 additions & 1 deletion conda/recipes/libcudf/meta.yaml
Original file line number Diff line number Diff line change
Expand Up @@ -247,7 +247,6 @@ outputs:
- test -f $PREFIX/include/cudf/structs/structs_column_view.hpp
- test -f $PREFIX/include/cudf/structs/struct_view.hpp
- test -f $PREFIX/include/cudf/structs/detail/concatenate.hpp
- test -f $PREFIX/include/cudf/structs/detail/contains.hpp
- test -f $PREFIX/include/cudf/table/table.hpp
- test -f $PREFIX/include/cudf/table/table_view.hpp
- test -f $PREFIX/include/cudf/tdigest/tdigest_column_view.cuh
Expand Down
2 changes: 1 addition & 1 deletion cpp/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -446,6 +446,7 @@ add_library(
src/scalar/scalar.cpp
src/scalar/scalar_factories.cpp
src/search/contains.cu
src/search/contains_nested.cu
src/search/search_ordered.cu
src/sort/is_sorted.cu
src/sort/rank.cu
Expand Down Expand Up @@ -514,7 +515,6 @@ add_library(
src/strings/utilities.cu
src/strings/wrap.cu
src/structs/copying/concatenate.cu
src/structs/search/contains.cu
src/structs/structs_column_factories.cu
src/structs/structs_column_view.cpp
src/structs/utilities.cpp
Expand Down
62 changes: 35 additions & 27 deletions cpp/include/cudf/detail/search.hpp
Original file line number Diff line number Diff line change
Expand Up @@ -23,35 +23,30 @@

#include <rmm/cuda_stream_view.hpp>

#include <vector>

namespace cudf {
namespace detail {
namespace cudf::detail {
/**
* @copydoc cudf::lower_bound
*
* @param stream CUDA stream used for device memory operations and kernel launches.
*/
std::unique_ptr<column> lower_bound(
table_view const& haystack,
table_view const& needles,
std::vector<order> const& column_order,
std::vector<null_order> const& null_precedence,
rmm::cuda_stream_view stream,
rmm::mr::device_memory_resource* mr = rmm::mr::get_current_device_resource());
std::unique_ptr<column> lower_bound(table_view const& haystack,
table_view const& needles,
std::vector<order> const& column_order,
std::vector<null_order> const& null_precedence,
rmm::cuda_stream_view stream,
rmm::mr::device_memory_resource* mr);

/**
* @copydoc cudf::upper_bound
*
* @param stream CUDA stream used for device memory operations and kernel launches.
*/
std::unique_ptr<column> upper_bound(
table_view const& haystack,
table_view const& needles,
std::vector<order> const& column_order,
std::vector<null_order> const& null_precedence,
rmm::cuda_stream_view stream,
rmm::mr::device_memory_resource* mr = rmm::mr::get_current_device_resource());
std::unique_ptr<column> upper_bound(table_view const& haystack,
table_view const& needles,
std::vector<order> const& column_order,
std::vector<null_order> const& null_precedence,
rmm::cuda_stream_view stream,
rmm::mr::device_memory_resource* mr);

/**
* @copydoc cudf::contains(column_view const&, scalar const&, rmm::mr::device_memory_resource*)
Expand All @@ -61,15 +56,28 @@ std::unique_ptr<column> upper_bound(
bool contains(column_view const& haystack, scalar const& needle, rmm::cuda_stream_view stream);

/**
* @copydoc cudf::contains(column_view const&, column_view const&, rmm::mr::device_memory_resource*)
*

PointKernel marked this conversation as resolved.
Show resolved Hide resolved
* @param stream CUDA stream used for device memory operations and kernel launches.
*/
std::unique_ptr<column> contains(
column_view const& haystack,
column_view const& needles,
rmm::cuda_stream_view stream,
rmm::mr::device_memory_resource* mr = rmm::mr::get_current_device_resource());
std::unique_ptr<column> contains(column_view const& haystack,
column_view const& needles,
rmm::cuda_stream_view stream,
rmm::mr::device_memory_resource* mr);

/**
* @brief Check if the (unique) row of the `needle` column is contained in the `haystack` column.
*
* If the input `needle` column has more than one row, only the first row will be considered.
*
* This function is designed for nested types only. It can also work with non-nested types
* but with lower performance due to the complexity of the implementation.
*
* @param haystack The column containing search space.
* @param needle A scalar value to check for existence in the search space.
* @return true if the given `needle` value exists in the `haystack` column.
*/
bool contains_nested_element(column_view const& haystack,
column_view const& needle,
rmm::cuda_stream_view stream);

} // namespace detail
} // namespace cudf
} // namespace cudf::detail
41 changes: 0 additions & 41 deletions cpp/include/cudf/structs/detail/contains.hpp

This file was deleted.

2 changes: 1 addition & 1 deletion cpp/src/hash/unordered_multiset.cuh
Original file line number Diff line number Diff line change
Expand Up @@ -131,7 +131,7 @@ class unordered_multiset {
return unordered_multiset(d_col.size(), std::move(hash_bins_start), std::move(hash_data));
}

unordered_multiset_device_view<Element, Hasher, Equality> to_device()
unordered_multiset_device_view<Element, Hasher, Equality> to_device() const
{
return unordered_multiset_device_view<Element, Hasher, Equality>(
size, hash_bins.data(), hash_data.data());
Expand Down
Loading