Skip to content

Commit

Permalink
Harden thrust algorithms against evil iterators that overload `operat…
Browse files Browse the repository at this point in the history
…or,` (#2349)

We need to guard against such iterators in libcu++, so our tests conventionallly contain iterators that delete `operator,`. To allow using thrust with such iterators we need to add the void casts
  • Loading branch information
miscco committed Sep 3, 2024
1 parent 457e4d7 commit 6b76188
Show file tree
Hide file tree
Showing 6 changed files with 11 additions and 11 deletions.
4 changes: 2 additions & 2 deletions thrust/thrust/system/detail/sequential/general_copy.h
Original file line number Diff line number Diff line change
Expand Up @@ -85,7 +85,7 @@ _CCCL_EXEC_CHECK_DISABLE
template <typename InputIterator, typename OutputIterator>
_CCCL_HOST_DEVICE OutputIterator general_copy(InputIterator first, InputIterator last, OutputIterator result)
{
for (; first != last; ++first, ++result)
for (; first != last; ++first, (void) ++result)
{
// gcc 4.2 crashes while instantiating iter_assign
#if defined(_CCCL_COMPILER_GCC) && (THRUST_GCC_VERSION < 40300)
Expand All @@ -102,7 +102,7 @@ _CCCL_EXEC_CHECK_DISABLE
template <typename InputIterator, typename Size, typename OutputIterator>
_CCCL_HOST_DEVICE OutputIterator general_copy_n(InputIterator first, Size n, OutputIterator result)
{
for (; n > Size(0); ++first, ++result, --n)
for (; n > Size(0); ++first, (void) ++result, (void) --n)
{
// gcc 4.2 crashes while instantiating iter_assign
#if defined(_CCCL_COMPILER_GCC) && (THRUST_GCC_VERSION < 40300)
Expand Down
6 changes: 3 additions & 3 deletions thrust/thrust/system/detail/sequential/partition.h
Original file line number Diff line number Diff line change
Expand Up @@ -210,7 +210,7 @@ _CCCL_HOST_DEVICE ForwardIterator stable_partition(
TempRange temp(exec, first, last);

InputIterator stencil_iter = stencil;
for (TempIterator iter = temp.begin(); iter != temp.end(); ++iter, ++stencil_iter)
for (TempIterator iter = temp.begin(); iter != temp.end(); ++iter, (void) ++stencil_iter)
{
if (wrapped_pred(*stencil_iter))
{
Expand All @@ -222,7 +222,7 @@ _CCCL_HOST_DEVICE ForwardIterator stable_partition(
ForwardIterator middle = first;
stencil_iter = stencil;

for (TempIterator iter = temp.begin(); iter != temp.end(); ++iter, ++stencil_iter)
for (TempIterator iter = temp.begin(); iter != temp.end(); ++iter, (void) ++stencil_iter)
{
if (!wrapped_pred(*stencil_iter))
{
Expand Down Expand Up @@ -287,7 +287,7 @@ _CCCL_HOST_DEVICE thrust::pair<OutputIterator1, OutputIterator2> stable_partitio
// wrap pred
thrust::detail::wrapped_function<Predicate, bool> wrapped_pred{pred};

for (; first != last; ++first, ++stencil)
for (; first != last; ++first, (void) ++stencil)
{
if (wrapped_pred(*stencil))
{
Expand Down
2 changes: 1 addition & 1 deletion thrust/thrust/system/detail/sequential/reduce_by_key.h
Original file line number Diff line number Diff line change
Expand Up @@ -66,7 +66,7 @@ _CCCL_HOST_DEVICE thrust::pair<OutputIterator1, OutputIterator2> reduce_by_key(
InputKeyType temp_key = *keys_first;
TemporaryType temp_value = *values_first;

for (++keys_first, ++values_first; keys_first != keys_last; ++keys_first, ++values_first)
for (++keys_first, ++values_first; keys_first != keys_last; ++keys_first, (void) ++values_first)
{
InputKeyType key = *keys_first;
InputValueType value = *values_first;
Expand Down
4 changes: 2 additions & 2 deletions thrust/thrust/system/detail/sequential/scan.h
Original file line number Diff line number Diff line change
Expand Up @@ -68,7 +68,7 @@ _CCCL_HOST_DEVICE OutputIterator inclusive_scan(

*result = *first;

for (++first, ++result; first != last; ++first, ++result)
for (++first, ++result; first != last; ++first, (void) ++result)
{
*result = sum = wrapped_binary_op(sum, *first);
}
Expand Down Expand Up @@ -144,7 +144,7 @@ _CCCL_HOST_DEVICE OutputIterator exclusive_scan(
*result = sum;
sum = binary_op(sum, tmp);

for (++first, ++result; first != last; ++first, ++result)
for (++first, ++result; first != last; ++first, (void) ++result)
{
tmp = *first;
*result = sum;
Expand Down
4 changes: 2 additions & 2 deletions thrust/thrust/system/detail/sequential/scan_by_key.h
Original file line number Diff line number Diff line change
Expand Up @@ -70,7 +70,7 @@ _CCCL_HOST_DEVICE OutputIterator inclusive_scan_by_key(

*result = prev_value;

for (++first1, ++first2, ++result; first1 != last1; ++first1, ++first2, ++result)
for (++first1, ++first2, ++result; first1 != last1; ++first1, (void) ++first2, (void) ++result)
{
KeyType key = *first1;

Expand Down Expand Up @@ -123,7 +123,7 @@ _CCCL_HOST_DEVICE OutputIterator exclusive_scan_by_key(

next = binary_op(next, temp_value);

for (++first1, ++first2, ++result; first1 != last1; ++first1, ++first2, ++result)
for (++first1, ++first2, ++result; first1 != last1; ++first1, (void) ++first2, (void) ++result)
{
KeyType key = *first1;

Expand Down
2 changes: 1 addition & 1 deletion thrust/thrust/system/detail/sequential/unique_by_key.h
Original file line number Diff line number Diff line change
Expand Up @@ -65,7 +65,7 @@ _CCCL_HOST_DEVICE thrust::pair<OutputIterator1, OutputIterator2> unique_by_key_c
InputKeyType temp_key = *keys_first;
OutputValueType temp_value = *values_first;

for (++keys_first, ++values_first; keys_first != keys_last; ++keys_first, ++values_first)
for (++keys_first, ++values_first; keys_first != keys_last; ++keys_first, (void) ++values_first)
{
InputKeyType key = *keys_first;
OutputValueType value = *values_first;
Expand Down

0 comments on commit 6b76188

Please sign in to comment.