Skip to content

Commit

Permalink
Use thrust::identity. (#357)
Browse files Browse the repository at this point in the history 
This PR adapts a few device lambdas to use `thrust::identity`. This
helps lift out a bit of the diff from #343.
  • Loading branch information
@bdice
bdice authored Aug 22, 2023
1 parent d101b4c commit 70a21da
Show file tree
Hide file tree
Showing 3 changed files with 20 additions and 36 deletions.
33 changes: 12 additions & 21 deletions tests/dynamic_map/erase_test.cu
View file
Original file line number Diff line number Diff line change
Expand Up @@ -20,6 +20,7 @@

#include <thrust/device_vector.h>
#include <thrust/execution_policy.h>
#include <thrust/functional.h>
#include <thrust/sequence.h>

#include <catch2/catch_template_test_macros.hpp>
Expand Down Expand Up @@ -62,9 +63,7 @@ TEMPLATE_TEST_CASE_SIG("erase key",
map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

// keys were actaully deleted
REQUIRE(cuco::test::none_of(d_keys_exist.begin(),
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::none_of(d_keys_exist.begin(), d_keys_exist.end(), thrust::identity{}));

// ensures that map is reusing deleted slots
map.insert(pairs_begin, pairs_begin + num_keys);
Expand All @@ -73,21 +72,17 @@ TEMPLATE_TEST_CASE_SIG("erase key",

map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

REQUIRE(cuco::test::all_of(d_keys_exist.begin(),
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::all_of(d_keys_exist.begin(), d_keys_exist.end(), thrust::identity{}));

// erase can act selectively
map.erase(d_keys.begin(), d_keys.begin() + num_keys / 2);
map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

REQUIRE(cuco::test::none_of(d_keys_exist.begin(),
d_keys_exist.begin() + num_keys / 2,
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::none_of(
d_keys_exist.begin(), d_keys_exist.begin() + num_keys / 2, thrust::identity{}));

REQUIRE(cuco::test::all_of(d_keys_exist.begin() + num_keys / 2,
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::all_of(
d_keys_exist.begin() + num_keys / 2, d_keys_exist.end(), thrust::identity{}));

// clear map
map.erase(d_keys.begin() + num_keys / 2, d_keys.end());
Expand Down Expand Up @@ -115,13 +110,11 @@ TEMPLATE_TEST_CASE_SIG("erase key",
map.erase(d_keys.begin(), d_keys.begin() + 2 * num_keys);
map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

REQUIRE(cuco::test::none_of(d_keys_exist.begin(),
d_keys_exist.begin() + 2 * num_keys,
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::none_of(
d_keys_exist.begin(), d_keys_exist.begin() + 2 * num_keys, thrust::identity{}));

REQUIRE(cuco::test::all_of(d_keys_exist.begin() + 2 * num_keys,
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::all_of(
d_keys_exist.begin() + 2 * num_keys, d_keys_exist.end(), thrust::identity{}));

REQUIRE(map.get_size() == 2 * num_keys);
// check that keys can be successfully deleted from all submaps (some will be unsuccessful
Expand All @@ -130,9 +123,7 @@ TEMPLATE_TEST_CASE_SIG("erase key",

map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

REQUIRE(cuco::test::none_of(d_keys_exist.begin(),
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::none_of(d_keys_exist.begin(), d_keys_exist.end(), thrust::identity{}));

REQUIRE(map.get_size() == 0);
}
Expand Down
19 changes: 7 additions & 12 deletions tests/static_map/erase_test.cu
View file
Original file line number Diff line number Diff line change
Expand Up @@ -20,6 +20,7 @@

#include <thrust/device_vector.h>
#include <thrust/execution_policy.h>
#include <thrust/functional.h>
#include <thrust/iterator/zip_iterator.h>
#include <thrust/sequence.h>
#include <thrust/tuple.h>
Expand Down Expand Up @@ -59,30 +60,24 @@ TEMPLATE_TEST_CASE_SIG("erase key", "", ((typename T), T), (int32_t), (int64_t))

map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

REQUIRE(cuco::test::none_of(d_keys_exist.begin(),
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::none_of(d_keys_exist.begin(), d_keys_exist.end(), thrust::identity{}));

map.insert(pairs_begin, pairs_begin + num_keys);

REQUIRE(map.get_size() == num_keys);

map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

REQUIRE(cuco::test::all_of(d_keys_exist.begin(),
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::all_of(d_keys_exist.begin(), d_keys_exist.end(), thrust::identity{}));

map.erase(d_keys.begin(), d_keys.begin() + num_keys / 2);
map.contains(d_keys.begin(), d_keys.end(), d_keys_exist.begin());

REQUIRE(cuco::test::none_of(d_keys_exist.begin(),
d_keys_exist.begin() + num_keys / 2,
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::none_of(
d_keys_exist.begin(), d_keys_exist.begin() + num_keys / 2, thrust::identity{}));

REQUIRE(cuco::test::all_of(d_keys_exist.begin() + num_keys / 2,
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::all_of(
d_keys_exist.begin() + num_keys / 2, d_keys_exist.end(), thrust::identity{}));

map.erase(d_keys.begin() + num_keys / 2, d_keys.end());
REQUIRE(map.get_size() == 0);
Expand Down
4 changes: 1 addition & 3 deletions tests/static_map/shared_memory_test.cu
View file
Original file line number Diff line number Diff line change
Expand Up @@ -148,9 +148,7 @@ TEMPLATE_TEST_CASE_SIG("Shared memory static map",
d_keys_exist.data().get(),
d_keys_and_values_correct.data().get());

REQUIRE(cuco::test::none_of(d_keys_exist.begin(),
d_keys_exist.end(),
[] __device__(const bool key_found) { return key_found; }));
REQUIRE(cuco::test::none_of(d_keys_exist.begin(), d_keys_exist.end(), thrust::identity{}));
}
}

Expand Down

0 comments on commit 70a21da

Please sign in to comment.