From 4376e787a728a9c7981890f7813e542e1a7e7d94 Mon Sep 17 00:00:00 2001 From: Daniel Juenger <2955913+sleeepyjack@users.noreply.github.com> Date: Tue, 18 Jul 2023 12:46:06 +0000 Subject: [PATCH 1/2] Add specialization for is_bitwise_comparable> --- include/cuco/detail/traits.hpp | 9 +++++++++ include/cuco/pair.cuh | 10 ++++++++++ include/cuco/utility/traits.hpp | 12 +++--------- 3 files changed, 22 insertions(+), 9 deletions(-) diff --git a/include/cuco/detail/traits.hpp b/include/cuco/detail/traits.hpp index 313f95430..87cb61c63 100644 --- a/include/cuco/detail/traits.hpp +++ b/include/cuco/detail/traits.hpp @@ -24,6 +24,15 @@ namespace cuco::detail { +template +struct is_bitwise_comparable_impl : std::false_type { +}; + +template +struct is_bitwise_comparable_impl>> + : std::true_type { +}; + template struct is_std_pair_like : cuda::std::false_type { }; diff --git a/include/cuco/pair.cuh b/include/cuco/pair.cuh index 0a804cc04..168cb7580 100644 --- a/include/cuco/pair.cuh +++ b/include/cuco/pair.cuh @@ -18,6 +18,7 @@ #include #include +#include #include #include @@ -141,6 +142,15 @@ template __host__ __device__ constexpr bool operator==(cuco::pair const& lhs, cuco::pair const& rhs) noexcept; +/** + * @brief A pair of bitwise comparable types is also bitwise comparable + */ +template +struct is_bitwise_comparable> + : std::integral_constant && is_bitwise_comparable_v> { +}; + } // namespace cuco #include diff --git a/include/cuco/utility/traits.hpp b/include/cuco/utility/traits.hpp index 1a6252dcb..bd2a187dc 100644 --- a/include/cuco/utility/traits.hpp +++ b/include/cuco/utility/traits.hpp @@ -16,8 +16,7 @@ #pragma once -#include -#include +#include #include @@ -36,15 +35,10 @@ namespace cuco { * if a `NaN` bit pattern were used as the empty sentinel value, it may not compare bitwise equal to * other `NaN` bit patterns. * + * @note By default, only types with unique object representations are allowed */ -template -struct is_bitwise_comparable : std::false_type { -}; - -/// By default, only types with unique object representations are allowed template -struct is_bitwise_comparable>> - : std::true_type { +struct is_bitwise_comparable : detail::is_bitwise_comparable_impl { }; template From bf3928457acf0193be2081621d645f623a1b6b54 Mon Sep 17 00:00:00 2001 From: Daniel Juenger <2955913+sleeepyjack@users.noreply.github.com> Date: Tue, 18 Jul 2023 17:04:50 +0000 Subject: [PATCH 2/2] Check for padding --- include/cuco/pair.cuh | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/include/cuco/pair.cuh b/include/cuco/pair.cuh index 168cb7580..0f1a863f5 100644 --- a/include/cuco/pair.cuh +++ b/include/cuco/pair.cuh @@ -148,7 +148,8 @@ __host__ __device__ constexpr bool operator==(cuco::pair const& lhs, template struct is_bitwise_comparable> : std::integral_constant && is_bitwise_comparable_v> { + is_bitwise_comparable_v && is_bitwise_comparable_v && + sizeof(pair) == sizeof(First) + sizeof(Second)> { }; } // namespace cuco