Merge branch 'main' into stf_fix_cudastf_debug

NVIDIA · Feb 26, 2025 · 1a0e755 · 1a0e755
2 parents 70461bb + cd6a090
commit 1a0e755
Show file tree

Hide file tree

Showing 140 changed files with 511 additions and 481 deletions.
diff --git a/thrust/examples/cuda/range_view.cu b/thrust/examples/cuda/range_view.cu
@@ -3,6 +3,8 @@
 #include <thrust/for_each.h>
 #include <thrust/iterator/counting_iterator.h>
 
+#include <cuda/std/iterator>
+
 #include <iostream>
 
 #include "../include/host_device.h"
@@ -20,10 +22,10 @@ class range_view
 {
 public:
   using iterator        = Iterator;
-  using value_type      = typename thrust::iterator_traits<iterator>::value_type;
-  using pointer         = typename thrust::iterator_traits<iterator>::pointer;
-  using difference_type = typename thrust::iterator_traits<iterator>::difference_type;
-  using reference       = typename thrust::iterator_traits<iterator>::reference;
+  using value_type      = typename cuda::std::iterator_traits<iterator>::value_type;
+  using pointer         = typename cuda::std::iterator_traits<iterator>::pointer;
+  using difference_type = typename cuda::std::iterator_traits<iterator>::difference_type;
+  using reference       = typename cuda::std::iterator_traits<iterator>::reference;
 
 private:
   const iterator first;

diff --git a/thrust/examples/expand.cu b/thrust/examples/expand.cu
@@ -20,7 +20,7 @@
 template <typename InputIterator1, typename InputIterator2, typename OutputIterator>
 OutputIterator expand(InputIterator1 first1, InputIterator1 last1, InputIterator2 first2, OutputIterator output)
 {
-  using difference_type = typename thrust::iterator_difference<InputIterator1>::type;
+  using difference_type = typename cuda::std::iterator_traits<InputIterator1>::difference_type;
 
   difference_type input_size  = thrust::distance(first1, last1);
   difference_type output_size = thrust::reduce(first1, last1);

diff --git a/thrust/examples/repeated_range.cu b/thrust/examples/repeated_range.cu
@@ -21,7 +21,7 @@ template <typename Iterator>
 class repeated_range
 {
 public:
-  using difference_type = typename thrust::iterator_difference<Iterator>::type;
+  using difference_type = typename cuda::std::iterator_traits<Iterator>::difference_type;
 
   struct repeat_functor
   {

diff --git a/thrust/examples/strided_range.cu b/thrust/examples/strided_range.cu
@@ -21,7 +21,7 @@ template <typename Iterator>
 class strided_range
 {
 public:
-  using difference_type = typename thrust::iterator_difference<Iterator>::type;
+  using difference_type = typename cuda::std::iterator_traits<Iterator>::difference_type;
 
   struct stride_functor
   {

diff --git a/thrust/examples/tiled_range.cu b/thrust/examples/tiled_range.cu
@@ -21,7 +21,7 @@ template <typename Iterator>
 class tiled_range
 {
 public:
-  using difference_type = typename thrust::iterator_difference<Iterator>::type;
+  using difference_type = typename cuda::std::iterator_traits<Iterator>::difference_type;
 
   struct tile_functor
   {

diff --git a/thrust/testing/counting_iterator.cu b/thrust/testing/counting_iterator.cu
@@ -235,7 +235,7 @@ DECLARE_UNITTEST(TestCountingIteratorLowerBound);
 void TestCountingIteratorDifference()
 {
   using Iterator   = thrust::counting_iterator<std::uint64_t>;
-  using Difference = thrust::iterator_difference<Iterator>::type;
+  using Difference = thrust::detail::it_difference_t<Iterator>;
 
   Difference diff = std::numeric_limits<std::uint32_t>::max() + 1;
 

diff --git a/thrust/testing/cuda/adjacent_difference.cu b/thrust/testing/cuda/adjacent_difference.cu
@@ -96,9 +96,9 @@ DECLARE_UNITTEST(TestAdjacentDifferenceCudaStreams);
 struct detect_wrong_difference
 {
   using difference_type   = void;
-  using value_type        = void;
+  using value_type        = long long;
   using pointer           = void;
-  using reference         = void;
+  using reference         = detect_wrong_difference;
   using iterator_category = ::cuda::std::output_iterator_tag;
 
   bool* flag;

diff --git a/thrust/testing/omp/reduce_intervals.cu b/thrust/testing/omp/reduce_intervals.cu
@@ -8,7 +8,7 @@
 template <typename InputIterator, typename OutputIterator, typename BinaryFunction, typename Decomposition>
 void reduce_intervals(InputIterator input, OutputIterator output, BinaryFunction binary_op, Decomposition decomp)
 {
-  using OutputType = typename thrust::iterator_value<OutputIterator>::type;
+  using OutputType = thrust::detail::it_value_t<OutputIterator>;
   using index_type = typename Decomposition::index_type;
 
   // wrap binary_op

diff --git a/thrust/testing/sort_permutation_iterator.cu b/thrust/testing/sort_permutation_iterator.cu
@@ -10,7 +10,7 @@ template <typename Iterator>
 class strided_range
 {
 public:
-  using difference_type = typename thrust::iterator_difference<Iterator>::type;
+  using difference_type = thrust::detail::it_difference_t<Iterator>;
 
   struct stride_functor
   {

diff --git a/thrust/testing/trivial_sequence.cu b/thrust/testing/trivial_sequence.cu
@@ -10,7 +10,7 @@ void test(Iterator first, Iterator last)
   using System = typename thrust::iterator_system<Iterator>::type;
   System system;
   thrust::detail::trivial_sequence<Iterator, System> ts(system, first, last);
-  using ValueType = typename thrust::iterator_traits<Iterator>::value_type;
+  using ValueType = typename ::cuda::std::iterator_traits<Iterator>::value_type;
 
   ASSERT_EQUAL_QUIET((ValueType) ts.begin()[0], ValueType(0, 11));
   ASSERT_EQUAL_QUIET((ValueType) ts.begin()[1], ValueType(2, 11));

diff --git a/thrust/testing/unique.cu b/thrust/testing/unique.cu
@@ -79,7 +79,7 @@ void TestUniqueCopyDispatchImplicit()
 DECLARE_UNITTEST(TestUniqueCopyDispatchImplicit);
 
 template <typename ForwardIterator>
-typename thrust::iterator_traits<ForwardIterator>::difference_type
+typename ::cuda::std::iterator_traits<ForwardIterator>::difference_type
 unique_count(my_system& system, ForwardIterator, ForwardIterator)
 {
   system.validate_dispatch();
@@ -98,7 +98,8 @@ void TestUniqueCountDispatchExplicit()
 DECLARE_UNITTEST(TestUniqueCountDispatchExplicit);
 
 template <typename ForwardIterator>
-typename thrust::iterator_traits<ForwardIterator>::difference_type unique_count(my_tag, ForwardIterator, ForwardIterator)
+typename ::cuda::std::iterator_traits<ForwardIterator>::difference_type
+unique_count(my_tag, ForwardIterator, ForwardIterator)
 {
   return 13;
 }

diff --git a/thrust/testing/unittest/assertions.h b/thrust/testing/unittest/assertions.h
@@ -409,8 +409,8 @@ void assert_equal(
   const std::string& filename = "unknown",
   int lineno                  = -1)
 {
-  using difference_type = typename THRUST_NS_QUALIFIER::iterator_difference<ForwardIterator1>::type;
-  using InputType       = typename THRUST_NS_QUALIFIER::iterator_value<ForwardIterator1>::type;
+  using difference_type = THRUST_NS_QUALIFIER::detail::it_difference_t<ForwardIterator1>;
+  using InputType       = THRUST_NS_QUALIFIER::detail::it_value_t<ForwardIterator1>;
 
   bool failure = false;
 
@@ -486,7 +486,7 @@ void assert_equal(
   const std::string& filename = "unknown",
   int lineno                  = -1)
 {
-  using InputType = typename THRUST_NS_QUALIFIER::iterator_traits<ForwardIterator1>::value_type;
+  using InputType = typename ::cuda::std::iterator_traits<ForwardIterator1>::value_type;
   assert_equal(first1, last1, first2, last2, THRUST_NS_QUALIFIER::equal_to<InputType>(), filename, lineno);
 }
 
@@ -501,7 +501,7 @@ void assert_almost_equal(
   const double a_tol          = DEFAULT_ABSOLUTE_TOL,
   const double r_tol          = DEFAULT_RELATIVE_TOL)
 {
-  using InputType = typename THRUST_NS_QUALIFIER::iterator_traits<ForwardIterator1>::value_type;
+  using InputType = typename ::cuda::std::iterator_traits<ForwardIterator1>::value_type;
   assert_equal(first1, last1, first2, last2, almost_equal_to<InputType>(a_tol, r_tol), filename, lineno);
 }
 

diff --git a/thrust/testing/zip_iterator.cu b/thrust/testing/zip_iterator.cu
@@ -109,7 +109,7 @@ struct TestZipIteratorReference
     using IteratorTuple1 = tuple<Iterator1, Iterator2>;
     using ZipIterator1   = zip_iterator<IteratorTuple1>;
 
-    using zip_iterator_reference_type1 = typename iterator_reference<ZipIterator1>::type;
+    using zip_iterator_reference_type1 = thrust::detail::it_reference_t<ZipIterator1>;
 
     host_vector<T> h_variable(1);
 
@@ -128,7 +128,7 @@ struct TestZipIteratorReference
     using IteratorTuple2 = tuple<Iterator3, Iterator4>;
     using ZipIterator2   = zip_iterator<IteratorTuple2>;
 
-    using zip_iterator_reference_type2 = typename iterator_reference<ZipIterator2>::type;
+    using zip_iterator_reference_type2 = thrust::detail::it_reference_t<ZipIterator2>;
 
     device_vector<T> d_variable(1);
 

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/count.h b/thrust/thrust/count.h
@@ -89,7 +89,7 @@ THRUST_NAMESPACE_BEGIN
  *  \see https://en.cppreference.com/w/cpp/algorithm/count
  */
 template <typename DerivedPolicy, typename InputIterator, typename EqualityComparable>
-_CCCL_HOST_DEVICE typename thrust::iterator_traits<InputIterator>::difference_type
+_CCCL_HOST_DEVICE thrust::detail::it_difference_t<InputIterator>
 count(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
       InputIterator first,
       InputIterator last,
@@ -131,7 +131,7 @@ count(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
  *  \see https://en.cppreference.com/w/cpp/algorithm/count
  */
 template <typename InputIterator, typename EqualityComparable>
-typename thrust::iterator_traits<InputIterator>::difference_type
+thrust::detail::it_difference_t<InputIterator>
 count(InputIterator first, InputIterator last, const EqualityComparable& value);
 
 /*! \p count_if finds the number of elements in <tt>[first,last)</tt> for which
@@ -184,7 +184,7 @@ count(InputIterator first, InputIterator last, const EqualityComparable& value);
  *  \see https://en.cppreference.com/w/cpp/algorithm/count
  */
 template <typename DerivedPolicy, typename InputIterator, typename Predicate>
-_CCCL_HOST_DEVICE typename thrust::iterator_traits<InputIterator>::difference_type
+_CCCL_HOST_DEVICE thrust::detail::it_difference_t<InputIterator>
 count_if(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
          InputIterator first,
          InputIterator last,
@@ -234,8 +234,7 @@ count_if(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
  *  \see https://en.cppreference.com/w/cpp/algorithm/count
  */
 template <typename InputIterator, typename Predicate>
-typename thrust::iterator_traits<InputIterator>::difference_type
-count_if(InputIterator first, InputIterator last, Predicate pred);
+thrust::detail::it_difference_t<InputIterator> count_if(InputIterator first, InputIterator last, Predicate pred);
 
 /*! \} // end counting
  *  \} // end reductions

diff --git a/thrust/thrust/detail/advance.inl b/thrust/thrust/detail/advance.inl
@@ -43,23 +43,24 @@ _CCCL_HOST_DEVICE void advance(InputIterator& i, Distance n)
 }
 
 template <typename InputIterator>
-_CCCL_HOST_DEVICE InputIterator next(InputIterator i, typename iterator_traits<InputIterator>::difference_type n = 1)
+_CCCL_HOST_DEVICE InputIterator next(InputIterator i, thrust::detail::it_difference_t<InputIterator> n = 1)
 {
   thrust::system::detail::generic::advance(i, n);
   return i;
 }
 
 template <typename BidirectionalIterator>
 _CCCL_HOST_DEVICE BidirectionalIterator
-prev(BidirectionalIterator i, typename iterator_traits<BidirectionalIterator>::difference_type n = 1)
+prev(BidirectionalIterator i, thrust::detail::it_difference_t<BidirectionalIterator> n = 1)
 {
   thrust::system::detail::generic::advance(i, -n);
   return i;
 }
 
+// FIXME(bgruber): what does this prevent against?
 template <typename BidirectionalIterator>
 _CCCL_HOST_DEVICE
-typename detail::disable_if<has_difference_type<iterator_traits<BidirectionalIterator>>::value,
+typename detail::disable_if<has_difference_type<::cuda::std::iterator_traits<BidirectionalIterator>>::value,
                             BidirectionalIterator>::type
 prev(BidirectionalIterator i, typename detail::pointer_traits<BidirectionalIterator>::difference_type n = 1)
 {

diff --git a/thrust/thrust/detail/allocator/copy_construct_range.inl b/thrust/thrust/detail/allocator/copy_construct_range.inl
@@ -100,12 +100,12 @@ _CCCL_HOST_DEVICE enable_if_convertible_t<FromSystem, ToSystem, Pointer> uniniti
   ZipIterator end   = begin;
 
   // get a zip_iterator pointing to the end
-  const typename thrust::iterator_difference<InputIterator>::type n = thrust::distance(first, last);
+  const thrust::detail::it_difference_t<InputIterator> n = thrust::distance(first, last);
   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/allocator/tagged_allocator.h b/thrust/thrust/detail/allocator/tagged_allocator.h
@@ -60,8 +60,8 @@ class tagged_allocator
   using value_type      = T;
   using pointer         = typename thrust::detail::pointer_traits<Pointer>::template rebind<T>::other;
   using const_pointer   = typename thrust::detail::pointer_traits<Pointer>::template rebind<const T>::other;
-  using reference       = typename thrust::iterator_reference<pointer>::type;
-  using const_reference = typename thrust::iterator_reference<const_pointer>::type;
+  using reference       = thrust::detail::it_reference_t<pointer>;
+  using const_reference = thrust::detail::it_reference_t<const_pointer>;
   using size_type       = std::size_t;
   using difference_type = typename thrust::detail::pointer_traits<pointer>::difference_type;
   using system_type     = Tag;

diff --git a/thrust/thrust/detail/count.h b/thrust/thrust/detail/count.h
@@ -30,26 +30,25 @@
 THRUST_NAMESPACE_BEGIN
 
 template <typename DerivedPolicy, typename InputIterator, typename EqualityComparable>
-_CCCL_HOST_DEVICE typename thrust::iterator_traits<InputIterator>::difference_type
+_CCCL_HOST_DEVICE thrust::detail::it_difference_t<InputIterator>
 count(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
       InputIterator first,
       InputIterator last,
       const EqualityComparable& value);
 
 template <typename DerivedPolicy, typename InputIterator, typename Predicate>
-_CCCL_HOST_DEVICE typename thrust::iterator_traits<InputIterator>::difference_type
+_CCCL_HOST_DEVICE thrust::detail::it_difference_t<InputIterator>
 count_if(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
          InputIterator first,
          InputIterator last,
          Predicate pred);
 
 template <typename InputIterator, typename EqualityComparable>
-typename thrust::iterator_traits<InputIterator>::difference_type
+thrust::detail::it_difference_t<InputIterator>
 count(InputIterator first, InputIterator last, const EqualityComparable& value);
 
 template <typename InputIterator, typename Predicate>
-typename thrust::iterator_traits<InputIterator>::difference_type
-count_if(InputIterator first, InputIterator last, Predicate pred);
+thrust::detail::it_difference_t<InputIterator> count_if(InputIterator first, InputIterator last, Predicate pred);
 
 THRUST_NAMESPACE_END
 

diff --git a/thrust/thrust/detail/count.inl b/thrust/thrust/detail/count.inl
@@ -35,7 +35,7 @@ THRUST_NAMESPACE_BEGIN
 
 _CCCL_EXEC_CHECK_DISABLE
 template <typename DerivedPolicy, typename InputIterator, typename EqualityComparable>
-_CCCL_HOST_DEVICE typename thrust::iterator_traits<InputIterator>::difference_type
+_CCCL_HOST_DEVICE thrust::detail::it_difference_t<InputIterator>
 count(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
       InputIterator first,
       InputIterator last,
@@ -47,7 +47,7 @@ count(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
 
 _CCCL_EXEC_CHECK_DISABLE
 template <typename DerivedPolicy, typename InputIterator, typename Predicate>
-_CCCL_HOST_DEVICE typename thrust::iterator_traits<InputIterator>::difference_type
+_CCCL_HOST_DEVICE thrust::detail::it_difference_t<InputIterator>
 count_if(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
          InputIterator first,
          InputIterator last,
@@ -58,7 +58,7 @@ count_if(const thrust::detail::execution_policy_base<DerivedPolicy>& exec,
 } // end count_if()
 
 template <typename InputIterator, typename EqualityComparable>
-typename thrust::iterator_traits<InputIterator>::difference_type
+thrust::detail::it_difference_t<InputIterator>
 count(InputIterator first, InputIterator last, const EqualityComparable& value)
 {
   using thrust::system::detail::generic::select_system;
@@ -71,8 +71,7 @@ count(InputIterator first, InputIterator last, const EqualityComparable& value)
 } // end count()
 
 template <typename InputIterator, typename Predicate>
-typename thrust::iterator_traits<InputIterator>::difference_type
-count_if(InputIterator first, InputIterator last, Predicate pred)
+thrust::detail::it_difference_t<InputIterator> count_if(InputIterator first, InputIterator last, Predicate pred)
 {
   using thrust::system::detail::generic::select_system;
 

diff --git a/thrust/thrust/detail/distance.inl b/thrust/thrust/detail/distance.inl
@@ -34,8 +34,7 @@ THRUST_NAMESPACE_BEGIN
 
 _CCCL_EXEC_CHECK_DISABLE
 template <typename InputIterator>
-inline _CCCL_HOST_DEVICE typename thrust::iterator_traits<InputIterator>::difference_type
-distance(InputIterator first, InputIterator last)
+inline _CCCL_HOST_DEVICE thrust::detail::it_difference_t<InputIterator> distance(InputIterator first, InputIterator last)
 {
   return thrust::system::detail::generic::distance(first, last);
 } // end distance()

diff --git a/thrust/thrust/detail/get_iterator_value.h b/thrust/thrust/detail/get_iterator_value.h
@@ -39,8 +39,7 @@ namespace detail
 // --------------------------------------------------
 // it is okay to dereference iterator in the usual way
 template <typename DerivedPolicy, typename Iterator>
-_CCCL_HOST_DEVICE typename thrust::iterator_traits<Iterator>::value_type
-get_iterator_value(thrust::execution_policy<DerivedPolicy>&, Iterator it)
+_CCCL_HOST_DEVICE it_value_t<Iterator> get_iterator_value(thrust::execution_policy<DerivedPolicy>&, Iterator it)
 {
   return *it;
 } // get_iterator_value(exec,Iterator);