#include #include #include #include #include #include #include #include #include #include // ensure that we properly support thrust::transform_iterator from cuda::std void TestTransformIteratorTraits() { using func = ::cuda::std::negate; using base_it = thrust::host_vector::iterator; using it = thrust::transform_iterator; using traits = cuda::std::iterator_traits; static_assert(cuda::std::is_same_v); static_assert(cuda::std::is_same_v); static_assert(cuda::std::is_same_v); static_assert(cuda::std::is_same_v); static_assert(cuda::std::is_same_v); static_assert(cuda::std::is_same_v, thrust::random_access_traversal_tag>); static_assert(cuda::std::__has_random_access_traversal); static_assert(!cuda::std::output_iterator); static_assert(cuda::std::input_iterator); static_assert(cuda::std::forward_iterator); static_assert(cuda::std::bidirectional_iterator); static_assert(cuda::std::random_access_iterator); static_assert(!cuda::std::contiguous_iterator); } DECLARE_UNITTEST(TestTransformIteratorTraits); template void TestTransformIterator() { using T = typename Vector::value_type; using UnaryFunction = ::cuda::std::negate; using Iterator = typename Vector::iterator; Vector input(4); Vector output(4); // initialize input thrust::sequence(input.begin(), input.end(), 1); // construct transform_iterator thrust::transform_iterator iter(input.begin(), UnaryFunction()); thrust::copy(iter, iter + 4, output.begin()); Vector ref{-1, -2, -3, -4}; ASSERT_EQUAL(output, ref); } DECLARE_VECTOR_UNITTEST(TestTransformIterator); template void TestMakeTransformIterator() { using T = typename Vector::value_type; using UnaryFunction = ::cuda::std::negate; using Iterator = typename Vector::iterator; Vector input(4); Vector output(4); // initialize input thrust::sequence(input.begin(), input.end(), 1); // construct transform_iterator thrust::transform_iterator iter(input.begin(), UnaryFunction()); thrust::copy(thrust::make_transform_iterator(input.begin(), UnaryFunction()), thrust::make_transform_iterator(input.end(), UnaryFunction()), output.begin()); Vector ref{-1, -2, -3, -4}; ASSERT_EQUAL(output, ref); } DECLARE_VECTOR_UNITTEST(TestMakeTransformIterator); template struct TestTransformIteratorReduce { void operator()(const size_t n) { thrust::host_vector h_data = unittest::random_samples(n); thrust::device_vector d_data = h_data; // run on host T h_result = thrust::reduce(thrust::make_transform_iterator(h_data.begin(), ::cuda::std::negate()), thrust::make_transform_iterator(h_data.end(), ::cuda::std::negate())); // run on device T d_result = thrust::reduce(thrust::make_transform_iterator(d_data.begin(), ::cuda::std::negate()), thrust::make_transform_iterator(d_data.end(), ::cuda::std::negate())); ASSERT_EQUAL(h_result, d_result); } }; VariableUnitTest TestTransformIteratorReduceInstance; struct ExtractValue { int operator()(std::unique_ptr const& n) { return *n; } }; void TestTransformIteratorNonCopyable() { thrust::host_vector> hv(4); hv[0] = std::make_unique(1); hv[1] = std::make_unique(2); hv[2] = std::make_unique(3); hv[3] = std::make_unique(4); auto transformed = thrust::make_transform_iterator(hv.begin(), ExtractValue{}); ASSERT_EQUAL(transformed[0], 1); ASSERT_EQUAL(transformed[1], 2); ASSERT_EQUAL(transformed[2], 3); ASSERT_EQUAL(transformed[3], 4); } DECLARE_UNITTEST(TestTransformIteratorNonCopyable); struct flip_value { _CCCL_HOST_DEVICE bool operator()(bool b) const { return !b; } }; struct pass_ref { _CCCL_HOST_DEVICE const bool& operator()(const bool& b _CCCL_LIFETIMEBOUND) const { return b; } }; // a user provided functor that forwards its argument struct forward { template constexpr _Tp&& operator()(_Tp&& __t) const noexcept { return ::cuda::std::forward<_Tp>(__t); } }; void TestTransformIteratorReferenceAndValueType() { using ::cuda::std::is_same; using ::cuda::std::negate; { thrust::host_vector v; auto it = v.begin(); static_assert(is_same::value); // ordinary reference static_assert(is_same::value); [[maybe_unused]] auto it_tr_val = thrust::make_transform_iterator(it, flip_value{}); static_assert(is_same::value); static_assert(is_same::value); [[maybe_unused]] auto it_tr_ref = thrust::make_transform_iterator(it, pass_ref{}); static_assert(is_same::value); static_assert(is_same::value); [[maybe_unused]] auto it_tr_fwd = thrust::make_transform_iterator(it, forward{}); static_assert(is_same::value); static_assert(is_same::value); [[maybe_unused]] auto it_tr_cid = thrust::make_transform_iterator(it, cuda::std::identity{}); static_assert(is_same::value); // special handling by // transform_iterator_reference static_assert(is_same::value); } { thrust::device_vector v; auto it = v.begin(); static_assert(is_same>::value); // proxy reference static_assert(is_same::value); [[maybe_unused]] auto it_tr_val = thrust::make_transform_iterator(it, flip_value{}); static_assert(is_same::value); static_assert(is_same::value); [[maybe_unused]] auto it_tr_ref = thrust::make_transform_iterator(it, pass_ref{}); static_assert(is_same::value); static_assert(is_same::value); [[maybe_unused]] auto it_tr_fwd = thrust::make_transform_iterator(it, forward{}); static_assert(is_same::value); // wrapped reference is decayed static_assert(is_same::value); [[maybe_unused]] auto it_tr_cid = thrust::make_transform_iterator(it, cuda::std::identity{}); static_assert(is_same::value); // special handling by // transform_iterator_reference static_assert(is_same::value); } { std::vector v; auto it = v.begin(); static_assert(is_same::reference>::value); // proxy reference static_assert(is_same::value); [[maybe_unused]] auto it_tr_val = thrust::make_transform_iterator(it, flip_value{}); static_assert(is_same::value); static_assert(is_same::value); [[maybe_unused]] auto it_tr_ref = thrust::make_transform_iterator(it, pass_ref{}); static_assert(is_same::value); static_assert(is_same::value); [[maybe_unused]] auto it_tr_fwd = thrust::make_transform_iterator(it, forward{}); static_assert(is_same::value); // proxy reference is decayed static_assert(is_same::value); [[maybe_unused]] auto it_tr_cid = thrust::make_transform_iterator(it, cuda::std::identity{}); static_assert(is_same::value); // special handling by // transform_iterator_reference static_assert(is_same::value); } } DECLARE_UNITTEST(TestTransformIteratorReferenceAndValueType); void TestTransformIteratorIdentity() { thrust::device_vector v(3, 42); ASSERT_EQUAL(*thrust::make_transform_iterator(v.begin(), cuda::std::identity{}), 42); using namespace thrust::placeholders; ASSERT_EQUAL(*thrust::make_transform_iterator(v.begin(), _1), 42); } DECLARE_UNITTEST(TestTransformIteratorIdentity);