-
Notifications
You must be signed in to change notification settings - Fork 921
Commit
This commit does not belong to any branch on this repository, and may belong to a fork outside of the repository.
Force Thrust to use 32-bit offset type. (#17523)
This fixes the patch we use for Thrust to always get a 32-bit offset type. The net effect of this patch is that we are behaving as if `THRUST_FORCE_32_BIT_OFFSET_TYPE` is set. This replaces a previous patch which I mistakenly did not update between CCCL 2.6.x testing and 2.7.0-rc2 testing. In the future we hope to configure this with CMake and drop the patches, but that will require us to use features from NVIDIA/cccl#2844 (which is not available in 2.7.0-rc2). Authors: - Bradley Dice (https://github.com/bdice) Approvers: - David Wendt (https://github.com/davidwendt) - Vyas Ramasubramani (https://github.com/vyasr) URL: #17523
- Loading branch information
Showing
1 changed file
with
19 additions
and
56 deletions.
There are no files selected for viewing
75 changes: 19 additions & 56 deletions
75
cpp/cmake/thirdparty/patches/thrust_disable_64bit_dispatching.diff
This file contains bidirectional Unicode text that may be interpreted or compiled differently than what appears below. To review, open the file in an editor that reveals hidden Unicode characters.
Learn more about bidirectional Unicode characters
| Original file line number | Diff line number | Diff line change |
|---|---|---|
| @@ -1,59 +1,22 @@ | ||
| diff --git a/thrust/thrust/system/cuda/detail/dispatch.h b/thrust/thrust/system/cuda/detail/dispatch.h | ||
| index 971b93d62..0d6b25b07 100644 | ||
| index 3d004aa55..71ce86bea 100644 | ||
| --- a/thrust/thrust/system/cuda/detail/dispatch.h | ||
| +++ b/thrust/thrust/system/cuda/detail/dispatch.h | ||
| @@ -36,16 +36,15 @@ | ||
| * that callables for both branches consist of the same tokens, and is intended to be used with Thrust-style dispatch | ||
| * interfaces, that always deduce the size type from the arguments. | ||
| */ | ||
| -#define THRUST_INDEX_TYPE_DISPATCH(status, call, count, arguments) \ | ||
| - if (count <= thrust::detail::integer_traits<std::int32_t>::const_max) \ | ||
| - { \ | ||
| - auto THRUST_PP_CAT2(count, _fixed) = static_cast<std::int32_t>(count); \ | ||
| - status = call arguments; \ | ||
| - } \ | ||
| - else \ | ||
| - { \ | ||
| - auto THRUST_PP_CAT2(count, _fixed) = static_cast<std::int64_t>(count); \ | ||
| - status = call arguments; \ | ||
| +#define THRUST_INDEX_TYPE_DISPATCH(status, call, count, arguments) \ | ||
| + if (count <= thrust::detail::integer_traits<std::int32_t>::const_max) \ | ||
| + { \ | ||
| + auto THRUST_PP_CAT2(count, _fixed) = static_cast<std::int32_t>(count); \ | ||
| + status = call arguments; \ | ||
| + } \ | ||
| + else \ | ||
| + { \ | ||
| + throw std::runtime_error("THRUST_INDEX_TYPE_DISPATCH 64-bit count is unsupported in libcudf"); \ | ||
| } | ||
|
|
||
| /** | ||
| @@ -55,18 +54,16 @@ | ||
| * | ||
| * This version of the macro supports providing two count variables, which is necessary for set algorithms. | ||
| */ | ||
| -#define THRUST_DOUBLE_INDEX_TYPE_DISPATCH(status, call, count1, count2, arguments) \ | ||
| - if (count1 + count2 <= thrust::detail::integer_traits<std::int32_t>::const_max) \ | ||
| - { \ | ||
| - auto THRUST_PP_CAT2(count1, _fixed) = static_cast<std::int32_t>(count1); \ | ||
| - auto THRUST_PP_CAT2(count2, _fixed) = static_cast<std::int32_t>(count2); \ | ||
| - status = call arguments; \ | ||
| - } \ | ||
| - else \ | ||
| - { \ | ||
| - auto THRUST_PP_CAT2(count1, _fixed) = static_cast<std::int64_t>(count1); \ | ||
| - auto THRUST_PP_CAT2(count2, _fixed) = static_cast<std::int64_t>(count2); \ | ||
| - status = call arguments; \ | ||
| +#define THRUST_DOUBLE_INDEX_TYPE_DISPATCH(status, call, count1, count2, arguments) \ | ||
| + if (count1 + count2 <= thrust::detail::integer_traits<std::int32_t>::const_max) \ | ||
| + { \ | ||
| + auto THRUST_PP_CAT2(count1, _fixed) = static_cast<std::int32_t>(count1); \ | ||
| + auto THRUST_PP_CAT2(count2, _fixed) = static_cast<std::int32_t>(count2); \ | ||
| + status = call arguments; \ | ||
| + } \ | ||
| + else \ | ||
| + { \ | ||
| + throw std::runtime_error("THRUST_DOUBLE_INDEX_TYPE_DISPATCH 64-bit count is unsupported in libcudf"); \ | ||
| } | ||
|
|
||
| /** | ||
| @@ -63,7 +63,7 @@ | ||
| _THRUST_INDEX_TYPE_DISPATCH_GUARD_UNDERFLOW(count1) \ | ||
| _THRUST_INDEX_TYPE_DISPATCH_GUARD_UNDERFLOW(count2) | ||
|
|
||
| -#if defined(THRUST_FORCE_64_BIT_OFFSET_TYPE) | ||
| +#if 0 | ||
| //! @brief Always dispatches to 64 bit offset version of an algorithm | ||
| # define THRUST_INDEX_TYPE_DISPATCH(status, call, count, arguments) \ | ||
| _THRUST_INDEX_TYPE_DISPATCH_GUARD_UNDERFLOW(count) \ | ||
| @@ -89,7 +89,7 @@ | ||
| _THRUST_INDEX_TYPE_DISPATCH_GUARD_UNDERFLOW(count) \ | ||
| _THRUST_INDEX_TYPE_DISPATCH(std::uint64_t, status, call_64, count, arguments) | ||
|
|
||
| -#elif defined(THRUST_FORCE_32_BIT_OFFSET_TYPE) | ||
| +#elif 1 | ||
|
|
||
| //! @brief Ensures that the size of the input does not overflow the offset type | ||
| # define _THRUST_INDEX_TYPE_DISPATCH_GUARD_OVERFLOW(index_type, count) \ |