From ba3c35260df43cd789be3615c31695acadb3c78e Mon Sep 17 00:00:00 2001
From: Daniel Juenger <2955913+sleeepyjack@users.noreply.github.com>
Date: Fri, 28 Jul 2023 21:10:44 +0000
Subject: [PATCH 1/2] Fix alignment issues in load_chunk

---
 include/cuco/detail/hash_functions/utils.cuh | 5 +++--
 1 file changed, 3 insertions(+), 2 deletions(-)

diff --git a/include/cuco/detail/hash_functions/utils.cuh b/include/cuco/detail/hash_functions/utils.cuh
index a50779f23..de8838f56 100644
--- a/include/cuco/detail/hash_functions/utils.cuh
+++ b/include/cuco/detail/hash_functions/utils.cuh
@@ -21,8 +21,9 @@ namespace cuco::detail {
 template <typename T, typename U, typename Extent>
 constexpr __host__ __device__ T load_chunk(U const* const data, Extent index) noexcept
 {
-  auto const chunks = reinterpret_cast<T const*>(data);
-  return chunks[index];
+  T chunk;
+  memcpy(&chunk, data + index, sizeof(T));
+  return chunk;
 }
 
 };  // namespace cuco::detail
\ No newline at end of file

From 419ecc00e00af4124e5dd9bd877ac29f5327fb0f Mon Sep 17 00:00:00 2001
From: Daniel Juenger <2955913+sleeepyjack@users.noreply.github.com>
Date: Fri, 28 Jul 2023 23:22:20 +0000
Subject: [PATCH 2/2] Cast to byte pointer

---
 include/cuco/detail/hash_functions/utils.cuh | 3 ++-
 1 file changed, 2 insertions(+), 1 deletion(-)

diff --git a/include/cuco/detail/hash_functions/utils.cuh b/include/cuco/detail/hash_functions/utils.cuh
index de8838f56..37e279ba7 100644
--- a/include/cuco/detail/hash_functions/utils.cuh
+++ b/include/cuco/detail/hash_functions/utils.cuh
@@ -21,8 +21,9 @@ namespace cuco::detail {
 template <typename T, typename U, typename Extent>
 constexpr __host__ __device__ T load_chunk(U const* const data, Extent index) noexcept
 {
+  auto const bytes = reinterpret_cast<std::byte const*>(data);
   T chunk;
-  memcpy(&chunk, data + index, sizeof(T));
+  memcpy(&chunk, bytes + index * sizeof(T), sizeof(T));
   return chunk;
 }