Replace direct uses of iterator_traits

NVIDIA · Feb 25, 2025 · 599c5fb · 599c5fb
1 parent cc70ec5
commit 599c5fb
Show file tree

Hide file tree

Showing 31 changed files with 77 additions and 78 deletions.
diff --git a/thrust/thrust/advance.h b/thrust/thrust/advance.h
@@ -101,7 +101,7 @@ template <typename InputIterator, typename Distance>
 _CCCL_HOST_DEVICE
 InputIterator next(
   InputIterator i
-, typename iterator_traits<InputIterator>::difference_type n = 1
+, typename ::cuda::std::iterator_traits<InputIterator>::difference_type n = 1
 );
 #endif
 
@@ -135,7 +135,7 @@ template <typename BidirectionalIterator, typename Distance>
 _CCCL_HOST_DEVICE
 BidirectionalIterator prev(
   BidirectionalIterator i
-, typename iterator_traits<BidirectionalIterator>::difference_type n = 1
+, typename ::cuda::std::iterator_traits<BidirectionalIterator>::difference_type n = 1
 );
 #endif
 

diff --git a/thrust/thrust/detail/allocator/copy_construct_range.inl b/thrust/thrust/detail/allocator/copy_construct_range.inl
@@ -104,8 +104,8 @@ _CCCL_HOST_DEVICE enable_if_convertible_t<FromSystem, ToSystem, Pointer> uniniti
   thrust::advance(end, n);
 
   // create a functor
-  using InputType  = typename iterator_traits<InputIterator>::value_type;
-  using OutputType = typename iterator_traits<Pointer>::value_type;
+  using InputType  = it_value_t<InputIterator>;
+  using OutputType = it_value_t<Pointer>;
 
   // do the for_each
   // note we use to_system to dispatch the for_each
@@ -135,8 +135,8 @@ _CCCL_HOST_DEVICE enable_if_convertible_t<FromSystem, ToSystem, Pointer> uniniti
   ZipIterator begin = thrust::make_zip_iterator(thrust::make_tuple(first, result));
 
   // create a functor
-  using InputType  = typename iterator_traits<InputIterator>::value_type;
-  using OutputType = typename iterator_traits<Pointer>::value_type;
+  using InputType  = it_value_t<InputIterator>;
+  using OutputType = it_value_t<Pointer>;
 
   // do the for_each_n
   // note we use to_system to dispatch the for_each_n

diff --git a/thrust/thrust/detail/memory_algorithms.h b/thrust/thrust/detail/memory_algorithms.h
@@ -64,7 +64,7 @@ _CCCL_HOST_DEVICE ForwardIt destroy(ForwardIt first, ForwardIt last) noexcept
 template <typename Allocator, typename ForwardIt>
 _CCCL_HOST_DEVICE ForwardIt destroy(Allocator const& alloc, ForwardIt first, ForwardIt last) noexcept
 {
-  using T      = typename iterator_traits<ForwardIt>::value_type;
+  using T      = detail::it_value_t<ForwardIt>;
   using traits = typename detail::allocator_traits<
     ::cuda::std::remove_cv_t<::cuda::std::remove_reference_t<Allocator>>>::template rebind_traits<T>::other;
 
@@ -92,7 +92,7 @@ _CCCL_HOST_DEVICE ForwardIt destroy_n(ForwardIt first, Size n) noexcept
 template <typename Allocator, typename ForwardIt, typename Size>
 _CCCL_HOST_DEVICE ForwardIt destroy_n(Allocator const& alloc, ForwardIt first, Size n) noexcept
 {
-  using T      = typename iterator_traits<ForwardIt>::value_type;
+  using T      = detail::it_value_t<ForwardIt>;
   using traits = typename detail::allocator_traits<
     ::cuda::std::remove_cv_t<::cuda::std::remove_reference_t<Allocator>>>::template rebind_traits<T>::other;
 
@@ -109,7 +109,7 @@ _CCCL_HOST_DEVICE ForwardIt destroy_n(Allocator const& alloc, ForwardIt first, S
 template <typename ForwardIt, typename... Args>
 _CCCL_HOST_DEVICE void uninitialized_construct(ForwardIt first, ForwardIt last, Args const&... args)
 {
-  using T = typename iterator_traits<ForwardIt>::value_type;
+  using T = detail::it_value_t<ForwardIt>;
 
   ForwardIt current = first;
 
@@ -132,7 +132,7 @@ _CCCL_HOST_DEVICE void uninitialized_construct(ForwardIt first, ForwardIt last,
 template <typename Allocator, typename ForwardIt, typename... Args>
 void uninitialized_construct_with_allocator(Allocator const& alloc, ForwardIt first, ForwardIt last, Args const&... args)
 {
-  using T      = typename iterator_traits<ForwardIt>::value_type;
+  using T      = detail::it_value_t<ForwardIt>;
   using traits = typename detail::allocator_traits<
     typename std::remove_cv<typename std::remove_reference<Allocator>::type>::type>::template rebind_traits<T>;
 
@@ -159,7 +159,7 @@ void uninitialized_construct_with_allocator(Allocator const& alloc, ForwardIt fi
 template <typename ForwardIt, typename Size, typename... Args>
 void uninitialized_construct_n(ForwardIt first, Size n, Args const&... args)
 {
-  using T = typename iterator_traits<ForwardIt>::value_type;
+  using T = detail::it_value_t<ForwardIt>;
 
   ForwardIt current = first;
 
@@ -182,7 +182,7 @@ void uninitialized_construct_n(ForwardIt first, Size n, Args const&... args)
 template <typename Allocator, typename ForwardIt, typename Size, typename... Args>
 void uninitialized_construct_n_with_allocator(Allocator const& alloc, ForwardIt first, Size n, Args const&... args)
 {
-  using T      = typename iterator_traits<ForwardIt>::value_type;
+  using T      = detail::it_value_t<ForwardIt>;
   using traits = typename detail::allocator_traits<
     typename std::remove_cv<typename std::remove_reference<Allocator>::type>::type>::template rebind_traits<T>;
 

diff --git a/thrust/thrust/iterator/transform_iterator.h b/thrust/thrust/iterator/transform_iterator.h
@@ -94,7 +94,7 @@ struct make_transform_iterator_base
                      Iterator,
                      value_type,
                      use_default,
-                     typename iterator_traits<Iterator>::iterator_category,
+                     typename ::cuda::std::iterator_traits<Iterator>::iterator_category,
                      reference>;
 };
 

diff --git a/thrust/thrust/system/cuda/detail/adjacent_difference.h b/thrust/thrust/system/cuda/detail/adjacent_difference.h
@@ -207,7 +207,7 @@ template <class Derived, class InputIt, class OutputIt>
 OutputIt _CCCL_HOST_DEVICE
 adjacent_difference(execution_policy<Derived>& policy, InputIt first, InputIt last, OutputIt result)
 {
-  using input_type = typename iterator_traits<InputIt>::value_type;
+  using input_type = thrust::detail::it_value_t<InputIt>;
   return cuda_cub::adjacent_difference(policy, first, last, result, minus<input_type>());
 }
 

diff --git a/thrust/thrust/system/cuda/detail/copy_if.h b/thrust/thrust/system/cuda/detail/copy_if.h
@@ -188,7 +188,7 @@ THRUST_RUNTIME_FUNCTION OutputIt copy_if(
   OutputIt output,
   Predicate predicate)
 {
-  using size_type = typename iterator_traits<InputIt>::difference_type;
+  using size_type = iterator_difference_t<InputIt>;
 
   size_type num_items       = static_cast<size_type>(thrust::distance(first, last));
   cudaError_t status        = cudaSuccess;

diff --git a/thrust/thrust/system/cuda/detail/core/util.h b/thrust/thrust/system/cuda/detail/core/util.h
@@ -485,7 +485,7 @@ struct get_arch<Plan<Arch>>
 // BlockLoad
 // -----------
 // a helper metaprogram that returns type of a block loader
-template <class PtxPlan, class It, class T = typename iterator_traits<It>::value_type>
+template <class PtxPlan, class It, class T = thrust::detail::it_value_t<It>>
 struct BlockLoad
 {
   using type = cub::BlockLoad<T, PtxPlan::BLOCK_THREADS, PtxPlan::ITEMS_PER_THREAD, PtxPlan::LOAD_ALGORITHM, 1, 1>;

diff --git a/thrust/thrust/system/cuda/detail/count.h b/thrust/thrust/system/cuda/detail/count.h
@@ -49,18 +49,18 @@ namespace cuda_cub
 {
 
 template <class Derived, class InputIt, class UnaryPred>
-typename iterator_traits<InputIt>::difference_type _CCCL_HOST_DEVICE
+thrust::detail::it_difference_t<InputIt> _CCCL_HOST_DEVICE
 count_if(execution_policy<Derived>& policy, InputIt first, InputIt last, UnaryPred unary_pred)
 {
-  using size_type       = typename iterator_traits<InputIt>::difference_type;
+  using size_type       = thrust::detail::it_difference_t<InputIt>;
   using flag_iterator_t = transform_iterator<UnaryPred, InputIt, size_type, size_type>;
 
   return cuda_cub::reduce_n(
     policy, flag_iterator_t(first, unary_pred), thrust::distance(first, last), size_type(0), plus<size_type>());
 }
 
 template <class Derived, class InputIt, class Value>
-typename iterator_traits<InputIt>::difference_type _CCCL_HOST_DEVICE
+thrust::detail::it_difference_t<InputIt> _CCCL_HOST_DEVICE
 count(execution_policy<Derived>& policy, InputIt first, InputIt last, Value const& value)
 {
   return cuda_cub::count_if(policy, first, last, thrust::detail::equal_to_value<Value>(value));

diff --git a/thrust/thrust/system/cuda/detail/extrema.h b/thrust/thrust/system/cuda/detail/extrema.h
@@ -369,8 +369,8 @@ element(execution_policy<Derived>& policy, ItemsIt first, ItemsIt last, BinaryPr
     return last;
   }
 
-  using InputType = typename iterator_traits<ItemsIt>::value_type;
-  using IndexType = typename iterator_traits<ItemsIt>::difference_type;
+  using InputType = thrust::detail::it_value_t<ItemsIt>;
+  using IndexType = thrust::detail::it_difference_t<ItemsIt>;
 
   IndexType num_items = static_cast<IndexType>(thrust::distance(first, last));
 
@@ -442,8 +442,7 @@ minmax_element(execution_policy<Derived>& policy, ItemsIt first, ItemsIt last, B
   }
 
   THRUST_CDP_DISPATCH(
-    (using InputType = typename iterator_traits<ItemsIt>::value_type;
-     using IndexType = typename iterator_traits<ItemsIt>::difference_type;
+    (using InputType = thrust::detail::it_value_t<ItemsIt>; using IndexType = thrust::detail::it_difference_t<ItemsIt>;
 
      const auto num_items = static_cast<IndexType>(thrust::distance(first, last));
 

diff --git a/thrust/thrust/system/cuda/detail/find.h b/thrust/thrust/system/cuda/detail/find.h
@@ -97,7 +97,7 @@ template <class ValueType, class InputIt, class UnaryOp>
 struct transform_input_iterator_t
 {
   using self_t            = transform_input_iterator_t;
-  using difference_type   = typename iterator_traits<InputIt>::difference_type;
+  using difference_type   = iterator_difference_t<InputIt>;
   using value_type        = ValueType;
   using pointer           = void;
   using reference         = value_type;

diff --git a/thrust/thrust/system/cuda/detail/for_each.h b/thrust/thrust/system/cuda/detail/for_each.h
@@ -72,7 +72,7 @@ Input THRUST_FUNCTION for_each_n(execution_policy<Derived>& policy, Input first,
 template <class Derived, class Input, class UnaryOp>
 Input THRUST_FUNCTION for_each(execution_policy<Derived>& policy, Input first, Input last, UnaryOp op)
 {
-  using size_type = typename iterator_traits<Input>::difference_type;
+  using size_type = thrust::detail::it_difference_t<Input>;
   size_type count = static_cast<size_type>(thrust::distance(first, last));
 
   return THRUST_NS_QUALIFIER::cuda_cub::for_each_n(policy, first, count, op);

diff --git a/thrust/thrust/system/cuda/detail/internal/copy_cross_system.h b/thrust/thrust/system/cuda/detail/internal/copy_cross_system.h
@@ -85,7 +85,7 @@ OutputIt _CCCL_HOST cross_system_copy_n(
   thrust::detail::true_type) // trivial copy
 
 {
-  using InputTy = typename iterator_traits<InputIt>::value_type;
+  using InputTy = thrust::detail::it_value_t<InputIt>;
   if (n > 0)
   {
     trivial_device_copy(

diff --git a/thrust/thrust/system/cuda/detail/merge.h b/thrust/thrust/system/cuda/detail/merge.h
@@ -64,7 +64,7 @@ merge(execution_policy<Derived>& policy,
       CompareOp compare_op = {})
 {
   THRUST_CDP_DISPATCH(
-    (using size_type         = typename iterator_traits<KeysIt1>::difference_type;
+    (using size_type         = thrust::detail::it_difference_t<KeysIt1>;
      const auto num_keys1    = static_cast<size_type>(thrust::distance(keys1_begin, keys1_end));
      const auto num_keys2    = static_cast<size_type>(thrust::distance(keys2_begin, keys2_end));
      const auto num_keys_out = num_keys1 + num_keys2;
@@ -151,7 +151,7 @@ pair<KeysOutputIt, ItemsOutputIt> _CCCL_HOST_DEVICE merge_by_key(
   CompareOp compare_op = {})
 {
   THRUST_CDP_DISPATCH(
-    (using size_type = typename iterator_traits<KeysIt1>::difference_type;
+    (using size_type = thrust::detail::it_difference_t<KeysIt1>;
 
      const auto num_keys1    = static_cast<size_type>(thrust::distance(keys1_begin, keys1_end));
      const auto num_keys2    = static_cast<size_type>(thrust::distance(keys2_begin, keys2_end));

diff --git a/thrust/thrust/system/cuda/detail/mismatch.h b/thrust/thrust/system/cuda/detail/mismatch.h
@@ -69,7 +69,7 @@ template <class ValueType, class InputIt1, class InputIt2, class BinaryOp>
 struct transform_pair_of_input_iterators_t
 {
   using self_t            = transform_pair_of_input_iterators_t;
-  using difference_type   = typename iterator_traits<InputIt1>::difference_type;
+  using difference_type   = thrust::detail::it_difference_t<InputIt1>;
   using value_type        = ValueType;
   using pointer           = void;
   using reference         = value_type;

diff --git a/thrust/thrust/system/cuda/detail/partition.h b/thrust/thrust/system/cuda/detail/partition.h
@@ -166,7 +166,7 @@ THRUST_RUNTIME_FUNCTION std::size_t partition(
   OutputIt output,
   Predicate predicate)
 {
-  using size_type = typename iterator_traits<InputIt>::difference_type;
+  using size_type = thrust::detail::it_difference_t<InputIt>;
 
   size_type num_items = thrust::distance(first, last);
   std::size_t num_selected{};
@@ -241,7 +241,7 @@ THRUST_RUNTIME_FUNCTION InputIt inplace_partition(
   }
 
   // Element type of the input iterator
-  using value_t         = typename iterator_traits<InputIt>::value_type;
+  using value_t         = thrust::detail::it_value_t<InputIt>;
   std::size_t num_items = static_cast<std::size_t>(thrust::distance(first, last));
 
   // Allocate temporary storage, which will serve as the input to the partition

diff --git a/thrust/thrust/system/cuda/detail/reduce.h b/thrust/thrust/system/cuda/detail/reduce.h
@@ -856,7 +856,7 @@ reduce_n(execution_policy<Derived>& policy, InputIt first, Size num_items, T ini
 template <class Derived, class InputIt, class T, class BinaryOp>
 _CCCL_HOST_DEVICE T reduce(execution_policy<Derived>& policy, InputIt first, InputIt last, T init, BinaryOp binary_op)
 {
-  using size_type = typename iterator_traits<InputIt>::difference_type;
+  using size_type = thrust::detail::it_difference_t<InputIt>;
   // FIXME: Check for RA iterator.
   size_type num_items = static_cast<size_type>(thrust::distance(first, last));
   return cuda_cub::reduce_n(policy, first, num_items, init, binary_op);
@@ -869,10 +869,10 @@ _CCCL_HOST_DEVICE T reduce(execution_policy<Derived>& policy, InputIt first, Inp
 }
 
 template <class Derived, class InputIt>
-_CCCL_HOST_DEVICE typename iterator_traits<InputIt>::value_type
+_CCCL_HOST_DEVICE thrust::detail::it_value_t<InputIt>
 reduce(execution_policy<Derived>& policy, InputIt first, InputIt last)
 {
-  using value_type = typename iterator_traits<InputIt>::value_type;
+  using value_type = thrust::detail::it_value_t<InputIt>;
   return cuda_cub::reduce(policy, first, last, value_type(0));
 }
 

diff --git a/thrust/thrust/system/cuda/detail/reduce_by_key.h b/thrust/thrust/system/cuda/detail/reduce_by_key.h
@@ -150,8 +150,8 @@ template <class KeysInputIt,
           class Size>
 struct ReduceByKeyAgent
 {
-  using key_type   = typename iterator_traits<KeysInputIt>::value_type;
-  using value_type = typename iterator_traits<ValuesInputIt>::value_type;
+  using key_type   = thrust::detail::it_value_t<KeysInputIt>;
+  using value_type = thrust::detail::it_value_t<ValuesInputIt>;
   using size_type  = Size;
 
   using size_value_pair_t = cub::KeyValuePair<size_type, value_type>;
@@ -893,7 +893,7 @@ THRUST_RUNTIME_FUNCTION pair<KeysOutputIt, ValuesOutputIt> reduce_by_key(
   EqualityOp equality_op,
   ReductionOp reduction_op)
 {
-  using size_type = typename iterator_traits<KeysInputIt>::difference_type;
+  using size_type = thrust::detail::it_difference_t<KeysInputIt>;
 
   size_type num_items = thrust::distance(keys_first, keys_last);
 

diff --git a/thrust/thrust/system/cuda/detail/replace.h b/thrust/thrust/system/cuda/detail/replace.h
@@ -127,7 +127,7 @@ OutputIt _CCCL_HOST_DEVICE replace_copy_if(
   Predicate predicate,
   T const& new_value)
 {
-  using output_type    = typename iterator_traits<OutputIt>::value_type;
+  using output_type    = thrust::detail::it_value_t<OutputIt>;
   using new_value_if_t = __replace::new_value_if_f<Predicate, T, output_type>;
   return cuda_cub::transform(policy, first, last, result, new_value_if_t(predicate, new_value));
 }
@@ -142,7 +142,7 @@ OutputIt _CCCL_HOST_DEVICE replace_copy_if(
   Predicate predicate,
   T const& new_value)
 {
-  using output_type    = typename iterator_traits<OutputIt>::value_type;
+  using output_type    = thrust::detail::it_value_t<OutputIt>;
   using new_value_if_t = __replace::new_value_if_f<Predicate, T, output_type>;
   return cuda_cub::transform(policy, first, last, stencil, result, new_value_if_t(predicate, new_value));
 }

diff --git a/thrust/thrust/system/cuda/detail/set_operations.h b/thrust/thrust/system/cuda/detail/set_operations.h
@@ -155,7 +155,7 @@ template <class It1, class It2, class Size, class Size2, class CompareOp>
 THRUST_DEVICE_FUNCTION pair<Size, Size>
 balanced_path(It1 keys1, It2 keys2, Size num_keys1, Size num_keys2, Size diag, Size2 levels, CompareOp compare_op)
 {
-  using T = typename iterator_traits<It1>::value_type;
+  using T = thrust::detail::it_value_t<It1>;
 
   Size index1 = merge_path<false>(keys1, num_keys1, keys2, num_keys2, diag, compare_op);
   Size index2 = diag - index1;
@@ -275,10 +275,10 @@ template <class KeysIt1,
           class HAS_VALUES>
 struct SetOpAgent
 {
-  using key1_type   = typename iterator_traits<KeysIt1>::value_type;
-  using key2_type   = typename iterator_traits<KeysIt2>::value_type;
-  using value1_type = typename iterator_traits<ValuesIt1>::value_type;
-  using value2_type = typename iterator_traits<ValuesIt2>::value_type;
+  using key1_type   = thrust::detail::it_value_t<KeysIt1>;
+  using key2_type   = thrust::detail::it_value_t<KeysIt2>;
+  using value1_type = thrust::detail::it_value_t<ValuesIt1>;
+  using value2_type = thrust::detail::it_value_t<ValuesIt2>;
 
   using key_type   = key1_type;
   using value_type = value1_type;
@@ -1149,7 +1149,7 @@ THRUST_RUNTIME_FUNCTION pair<KeysOutputIt, ValuesOutputIt> set_operations(
   CompareOp compare_op,
   SetOp set_op)
 {
-  using size_type = typename iterator_traits<KeysIt1>::difference_type;
+  using size_type = thrust::detail::it_difference_t<KeysIt1>;
 
   size_type num_keys1 = static_cast<size_type>(thrust::distance(keys1_first, keys1_last));
   size_type num_keys2 = static_cast<size_type>(thrust::distance(keys2_first, keys2_last));

diff --git a/thrust/thrust/system/cuda/detail/sort.h b/thrust/thrust/system/cuda/detail/sort.h
@@ -160,7 +160,7 @@ THRUST_RUNTIME_FUNCTION void merge_sort(
   execution_policy<Derived>& policy, KeysIt keys_first, KeysIt keys_last, ItemsIt items_first, CompareOp compare_op)
 
 {
-  using size_type = typename iterator_traits<KeysIt>::difference_type;
+  using size_type = thrust::detail::it_difference_t<KeysIt>;
 
   size_type count = static_cast<size_type>(thrust::distance(keys_first, keys_last));
 

diff --git a/thrust/thrust/system/cuda/detail/swap_ranges.h b/thrust/thrust/system/cuda/detail/swap_ranges.h
@@ -61,8 +61,8 @@ struct swap_f
   ItemsIt1 items1;
   ItemsIt2 items2;
 
-  using value1_type = typename iterator_traits<ItemsIt1>::value_type;
-  using value2_type = typename iterator_traits<ItemsIt2>::value_type;
+  using value1_type = thrust::detail::it_value_t<ItemsIt1>;
+  using value2_type = thrust::detail::it_value_t<ItemsIt2>;
 
   THRUST_FUNCTION
   swap_f(ItemsIt1 items1_, ItemsIt2 items2_)
@@ -88,7 +88,7 @@ template <class Derived, class ItemsIt1, class ItemsIt2>
 ItemsIt2 _CCCL_HOST_DEVICE
 swap_ranges(execution_policy<Derived>& policy, ItemsIt1 first1, ItemsIt1 last1, ItemsIt2 first2)
 {
-  using size_type = typename iterator_traits<ItemsIt1>::difference_type;
+  using size_type = thrust::detail::it_difference_t<ItemsIt1>;
 
   size_type num_items = static_cast<size_type>(thrust::distance(first1, last1));
 

diff --git a/thrust/thrust/system/cuda/detail/tabulate.h b/thrust/thrust/system/cuda/detail/tabulate.h
@@ -72,7 +72,7 @@ struct functor
 template <class Derived, class Iterator, class TabulateOp>
 void _CCCL_HOST_DEVICE tabulate(execution_policy<Derived>& policy, Iterator first, Iterator last, TabulateOp tabulate_op)
 {
-  using size_type = typename iterator_traits<Iterator>::difference_type;
+  using size_type = thrust::detail::it_difference_t<Iterator>;
 
   size_type count = thrust::distance(first, last);