diff options
author | Michael Schellenberger Costa <miscco@nvidia.com> | 2022-09-19 21:17:31 +0300 |
---|---|---|
committer | Michael Schellenberger Costa <miscco@nvidia.com> | 2022-09-19 21:17:31 +0300 |
commit | 9a088f27620273d801d2dfb82c9d43c282c4e0ca (patch) | |
tree | 6b095329b5d4fe3275ac144898b1cf1509d3799d | |
parent | 5029d334c334ff22c680bf3804040a105b6e89bc (diff) |
Replace `thrust::tuple` implementation with `cuda::std::tuple`replace_tuple
Our old implementation of `tuple` has a lot of limitations.
So rather than using the old homegrown version reuse `cuda::std::tuple`
-rw-r--r-- | testing/tuple_sort.cu | 5 | ||||
-rw-r--r-- | testing/tuple_transform.cu | 5 | ||||
-rw-r--r-- | testing/zip_iterator.cu | 6 | ||||
-rw-r--r-- | thrust/detail/functional/actor.h | 6 | ||||
-rw-r--r-- | thrust/detail/functional/actor.inl | 12 | ||||
-rw-r--r-- | thrust/detail/functional/argument.h | 6 | ||||
-rw-r--r-- | thrust/detail/functional/composite.h | 41 | ||||
-rw-r--r-- | thrust/detail/functional/operators/operator_adaptors.h | 20 | ||||
-rw-r--r-- | thrust/detail/pair.inl | 231 | ||||
-rw-r--r-- | thrust/detail/tuple.inl | 1004 | ||||
-rw-r--r-- | thrust/iterator/detail/tuple_of_iterator_references.h | 174 | ||||
-rw-r--r-- | thrust/pair.h | 320 | ||||
-rw-r--r-- | thrust/tuple.h | 476 |
13 files changed, 225 insertions, 2081 deletions
diff --git a/testing/tuple_sort.cu b/testing/tuple_sort.cu index 1bad8b86..227f3412 100644 --- a/testing/tuple_sort.cu +++ b/testing/tuple_sort.cu @@ -20,10 +20,7 @@ struct GetFunctor { template<typename Tuple> THRUST_HOST_DEVICE - typename thrust::access_traits< - typename thrust::tuple_element<N, Tuple>::type - >::const_type - operator()(const Tuple &t) + decltype(auto) operator()(const Tuple &t) { return thrust::get<N>(t); } diff --git a/testing/tuple_transform.cu b/testing/tuple_transform.cu index 43ebeac9..669db43f 100644 --- a/testing/tuple_transform.cu +++ b/testing/tuple_transform.cu @@ -19,10 +19,7 @@ struct GetFunctor { template<typename Tuple> THRUST_HOST_DEVICE - typename thrust::access_traits< - typename thrust::tuple_element<N, Tuple>::type - >::const_type - operator()(const Tuple &t) + decltype(auto) operator()(const Tuple &t) { return thrust::get<N>(t); } diff --git a/testing/zip_iterator.cu b/testing/zip_iterator.cu index e5329fbc..09cbdcd2 100644 --- a/testing/zip_iterator.cu +++ b/testing/zip_iterator.cu @@ -285,9 +285,9 @@ void TestZipIteratorCopy(void) sequence(input0.begin(), input0.end(), T{0}); sequence(input1.begin(), input1.end(), T{13}); - copy( make_zip_iterator(make_tuple(input0.begin(), input1.begin())), - make_zip_iterator(make_tuple(input0.end(), input1.end())), - make_zip_iterator(make_tuple(output0.begin(), output1.begin()))); + thrust::copy( make_zip_iterator(make_tuple(input0.begin(), input1.begin())), + make_zip_iterator(make_tuple(input0.end(), input1.end())), + make_zip_iterator(make_tuple(output0.begin(), output1.begin()))); ASSERT_EQUAL(input0, output0); ASSERT_EQUAL(input1, output1); diff --git a/thrust/detail/functional/actor.h b/thrust/detail/functional/actor.h index b677e3f0..b40a3c25 100644 --- a/thrust/detail/functional/actor.h +++ b/thrust/detail/functional/actor.h @@ -65,10 +65,6 @@ template<typename Eval> THRUST_HOST_DEVICE actor(const Eval &base); - THRUST_HOST_DEVICE - typename apply_actor<eval_type, thrust::null_type >::type - operator()(void) const; - template <typename... Ts> THRUST_HOST_DEVICE typename apply_actor<eval_type, thrust::tuple<eval_ref<Ts>...>>::type @@ -122,7 +118,7 @@ template<typename Eval> { typedef typename thrust::detail::functional::apply_actor< thrust::detail::functional::actor<Eval>, - thrust::null_type + thrust::tuple<> >::type type; }; // end result_of diff --git a/thrust/detail/functional/actor.inl b/thrust/detail/functional/actor.inl index 2f2c7bea..8fca46f5 100644 --- a/thrust/detail/functional/actor.inl +++ b/thrust/detail/functional/actor.inl @@ -54,18 +54,6 @@ template<typename Eval> : eval_type(base) {} -template<typename Eval> - THRUST_HOST_DEVICE - typename apply_actor< - typename actor<Eval>::eval_type, - typename thrust::null_type - >::type - actor<Eval> - ::operator()(void) const -{ - return eval_type::eval(thrust::null_type()); -} // end basic_environment::operator() - // actor::operator() needs to construct a tuple of references to its // arguments. To make this work with thrust::reference<T>, we need to // detect thrust proxy references and store them as T rather than T&. diff --git a/thrust/detail/functional/argument.h b/thrust/detail/functional/argument.h index c9479ab2..a34e07a7 100644 --- a/thrust/detail/functional/argument.h +++ b/thrust/detail/functional/argument.h @@ -41,9 +41,9 @@ template<unsigned int i, typename Env> }; template<unsigned int i> - struct argument_helper<i,thrust::null_type> + struct argument_helper<i,thrust::tuple<>> { - typedef thrust::null_type type; + typedef thrust::tuple<> type; }; @@ -52,7 +52,7 @@ template<unsigned int i> { public: template<typename Env> - struct result + struct result : argument_helper<i,Env> { }; diff --git a/thrust/detail/functional/composite.h b/thrust/detail/functional/composite.h index 35f34f66..14e14244 100644 --- a/thrust/detail/functional/composite.h +++ b/thrust/detail/functional/composite.h @@ -36,33 +36,11 @@ namespace detail namespace functional { -// XXX we should just take a single EvalTuple -template<typename Eval0, - typename Eval1 = thrust::null_type, - typename Eval2 = thrust::null_type, - typename Eval3 = thrust::null_type, - typename Eval4 = thrust::null_type, - typename Eval5 = thrust::null_type, - typename Eval6 = thrust::null_type, - typename Eval7 = thrust::null_type, - typename Eval8 = thrust::null_type, - typename Eval9 = thrust::null_type, - typename Eval10 = thrust::null_type> - class composite; +template <typename... Eval> +class composite; template<typename Eval0, typename Eval1> - class composite< - Eval0, - Eval1, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type - > + class composite<Eval0, Eval1> { public: template<typename Env> @@ -96,18 +74,7 @@ template<typename Eval0, typename Eval1> }; // end composite<Eval0,Eval1> template<typename Eval0, typename Eval1, typename Eval2> - class composite< - Eval0, - Eval1, - Eval2, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type, - thrust::null_type - > + class composite<Eval0, Eval1, Eval2> { public: template<typename Env> diff --git a/thrust/detail/functional/operators/operator_adaptors.h b/thrust/detail/functional/operators/operator_adaptors.h index 9321fccc..59d340b2 100644 --- a/thrust/detail/functional/operators/operator_adaptors.h +++ b/thrust/detail/functional/operators/operator_adaptors.h @@ -43,7 +43,7 @@ struct transparent_unary_operator using argument = typename thrust::detail::eval_if< thrust::tuple_size<Env>::value != 1, - thrust::detail::identity_<thrust::null_type>, + thrust::detail::identity_<thrust::tuple<>>, thrust::detail::functional::argument_helper<0, Env> >::type; @@ -57,8 +57,8 @@ struct transparent_unary_operator template <typename Env> using result_type = typename thrust::detail::eval_if< - std::is_same<thrust::null_type, argument<Env>>::value, - thrust::detail::identity_<thrust::null_type>, + std::is_same<thrust::tuple<>, argument<Env>>::value, + thrust::detail::identity_<thrust::tuple<>>, result_type_impl<Env> >::type; @@ -88,16 +88,16 @@ struct transparent_binary_operator using first_argument = typename thrust::detail::eval_if< thrust::tuple_size<Env>::value != 2, - thrust::detail::identity_<thrust::null_type>, - thrust::detail::functional::argument_helper<0, Env> + thrust::detail::identity_<thrust::tuple<>>, + thrust::detail::functional::argument_helper<0, Env> >::type; template <typename Env> using second_argument = typename thrust::detail::eval_if< thrust::tuple_size<Env>::value != 2, - thrust::detail::identity_<thrust::null_type>, - thrust::detail::functional::argument_helper<1, Env> + thrust::detail::identity_<thrust::tuple<>>, + thrust::detail::functional::argument_helper<1, Env> >::type; template <typename Env> @@ -111,9 +111,9 @@ struct transparent_binary_operator template <typename Env> using result_type = typename thrust::detail::eval_if< - (std::is_same<thrust::null_type, first_argument<Env>>::value || - std::is_same<thrust::null_type, second_argument<Env>>::value), - thrust::detail::identity_<thrust::null_type>, + (std::is_same<thrust::tuple<>, first_argument<Env>>::value || + std::is_same<thrust::tuple<>, second_argument<Env>>::value), + thrust::detail::identity_<thrust::tuple<>>, result_type_impl<Env> >::type; diff --git a/thrust/detail/pair.inl b/thrust/detail/pair.inl deleted file mode 100644 index 99118656..00000000 --- a/thrust/detail/pair.inl +++ /dev/null @@ -1,231 +0,0 @@ -/* - * Copyright 2008-2021 NVIDIA Corporation - * - * Licensed under the Apache License, Version 2.0 (the "License"); - * you may not use this file except in compliance with the License. - * You may obtain a copy of the License at - * - * http://www.apache.org/licenses/LICENSE-2.0 - * - * Unless required by applicable law or agreed to in writing, software - * distributed under the License is distributed on an "AS IS" BASIS, - * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. - * See the License for the specific language governing permissions and - * limitations under the License. - */ - -#pragma once - -#include <thrust/detail/config.h> - -#include <thrust/pair.h> -#include <thrust/detail/swap.h> -#include <thrust/tuple.h> - -THRUST_NAMESPACE_BEGIN - -template <typename T1, typename T2> - THRUST_HOST_DEVICE - pair<T1,T2> - ::pair(void) - :first(),second() -{ - ; -} // end pair::pair() - - -template <typename T1, typename T2> - THRUST_HOST_DEVICE - pair<T1,T2> - ::pair(const T1 &x, const T2 &y) - :first(x),second(y) -{ - ; -} // end pair::pair() - - -template <typename T1, typename T2> - template <typename U1, typename U2> - THRUST_HOST_DEVICE - pair<T1,T2> - ::pair(const pair<U1,U2> &p) - :first(p.first),second(p.second) -{ - ; -} // end pair::pair() - - -template <typename T1, typename T2> - template <typename U1, typename U2> - THRUST_HOST_DEVICE - pair<T1,T2> - ::pair(const std::pair<U1,U2> &p) - :first(p.first),second(p.second) -{ - ; -} // end pair::pair() - - -template<typename T1, typename T2> - inline THRUST_HOST_DEVICE - void pair<T1,T2> - ::swap(thrust::pair<T1,T2> &p) -{ - using thrust::swap; - - swap(first, p.first); - swap(second, p.second); -} // end pair::swap() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator==(const pair<T1,T2> &x, const pair<T1,T2> &y) -{ - return x.first == y.first && x.second == y.second; -} // end operator==() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator<(const pair<T1,T2> &x, const pair<T1,T2> &y) -{ - return x.first < y.first || (!(y.first < x.first) && x.second < y.second); -} // end operator<() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator!=(const pair<T1,T2> &x, const pair<T1,T2> &y) -{ - return !(x == y); -} // end operator==() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator>(const pair<T1,T2> &x, const pair<T1,T2> &y) -{ - return y < x; -} // end operator<() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator<=(const pair<T1,T2> &x, const pair<T1,T2> &y) -{ - return !(y < x); -} // end operator<=() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator>=(const pair<T1,T2> &x, const pair<T1,T2> &y) -{ - return !(x < y); -} // end operator>=() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - void swap(pair<T1,T2> &x, pair<T1,T2> &y) -{ - return x.swap(y); -} // end swap() - - -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - pair<T1,T2> make_pair(T1 x, T2 y) -{ - return pair<T1,T2>(x,y); -} // end make_pair() - - -// specializations of tuple_element for pair -template<typename T1, typename T2> - struct tuple_element<0, pair<T1,T2>> -{ - typedef T1 type; -}; // end tuple_element - -template<typename T1, typename T2> - struct tuple_element<1, pair<T1,T2>> -{ - typedef T2 type; -}; // end tuple_element - - -// specialization of tuple_size for pair -template<typename T1, typename T2> - struct tuple_size<pair<T1,T2>> -{ - static const unsigned int value = 2; -}; // end tuple_size - - - -namespace detail -{ - - -template<int N, typename Pair> struct pair_get {}; - -template<typename Pair> - struct pair_get<0, Pair> -{ - inline THRUST_HOST_DEVICE - const typename tuple_element<0, Pair>::type & - operator()(const Pair &p) const - { - return p.first; - } // end operator()() - - inline THRUST_HOST_DEVICE - typename tuple_element<0, Pair>::type & - operator()(Pair &p) const - { - return p.first; - } // end operator()() -}; // end pair_get - - -template<typename Pair> - struct pair_get<1, Pair> -{ - inline THRUST_HOST_DEVICE - const typename tuple_element<1, Pair>::type & - operator()(const Pair &p) const - { - return p.second; - } // end operator()() - - inline THRUST_HOST_DEVICE - typename tuple_element<1, Pair>::type & - operator()(Pair &p) const - { - return p.second; - } // end operator()() -}; // end pair_get - -} // end detail - - - -template<unsigned int N, typename T1, typename T2> - inline THRUST_HOST_DEVICE - typename tuple_element<N, pair<T1,T2> >::type & - get(pair<T1,T2> &p) -{ - return detail::pair_get<N, pair<T1,T2> >()(p); -} // end get() - -template<unsigned int N, typename T1, typename T2> - inline THRUST_HOST_DEVICE - const typename tuple_element<N, pair<T1,T2> >::type & - get(const pair<T1,T2> &p) -{ - return detail::pair_get<N, pair<T1,T2> >()(p); -} // end get() - -THRUST_NAMESPACE_END diff --git a/thrust/detail/tuple.inl b/thrust/detail/tuple.inl deleted file mode 100644 index 34434fb1..00000000 --- a/thrust/detail/tuple.inl +++ /dev/null @@ -1,1004 +0,0 @@ -/* - * Copyright 2008-2021 NVIDIA Corporation - * - * Licensed under the Apache License, Version 2.0 (the "License"); - * you may not use this file except in compliance with the License. - * You may obtain a copy of the License at - * - * http://www.apache.org/licenses/LICENSE-2.0 - * - * Unless required by applicable law or agreed to in writing, software - * distributed under the License is distributed on an "AS IS" BASIS, - * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. - * See the License for the specific language governing permissions and - * limitations under the License. - */ - -#pragma once - -#include <thrust/detail/config.h> - -#include <thrust/detail/type_traits.h> -#include <thrust/detail/swap.h> - -THRUST_NAMESPACE_BEGIN - -// define null_type -struct null_type {}; - -// null_type comparisons -THRUST_HOST_DEVICE inline -bool operator==(const null_type&, const null_type&) { return true; } - -THRUST_HOST_DEVICE inline -bool operator>=(const null_type&, const null_type&) { return true; } - -THRUST_HOST_DEVICE inline -bool operator<=(const null_type&, const null_type&) { return true; } - -THRUST_HOST_DEVICE inline -bool operator!=(const null_type&, const null_type&) { return false; } - -THRUST_HOST_DEVICE inline -bool operator<(const null_type&, const null_type&) { return false; } - -THRUST_HOST_DEVICE inline -bool operator>(const null_type&, const null_type&) { return false; } - -// forward declaration for tuple -template < - class T0 = null_type, class T1 = null_type, class T2 = null_type, - class T3 = null_type, class T4 = null_type, class T5 = null_type, - class T6 = null_type, class T7 = null_type, class T8 = null_type, - class T9 = null_type> -class tuple; - - -template <size_t N, class T> struct tuple_element; - -template<size_t N, class T> - struct tuple_element_impl -{ - private: - typedef typename T::tail_type Next; - - public: - /*! The result of this metafunction is returned in \c type. - */ - typedef typename tuple_element_impl<N-1, Next>::type type; -}; // end tuple_element - -template<class T> - struct tuple_element_impl<0,T> -{ - typedef typename T::head_type type; -}; - -template <size_t N, class T> - struct tuple_element<N, T const> -{ - using type = typename std::add_const<typename tuple_element<N, T>::type>::type; -}; - -template <size_t N, class T> -struct tuple_element<N, T volatile> -{ - using type = typename std::add_volatile<typename tuple_element<N, T>::type>::type; -}; - -template <size_t N, class T> - struct tuple_element<N, T const volatile> -{ - using type = typename std::add_cv<typename tuple_element<N, T>::type>::type; -}; - -template <size_t N, class T> -struct tuple_element{ - using type = typename tuple_element_impl<N,T>::type; -}; - -// forward declaration of tuple_size -template<class T> struct tuple_size; - -template<class T> - struct tuple_size<T const> : public tuple_size<T> {}; - -template<class T> - struct tuple_size<T volatile> : public tuple_size<T> {}; - -template<class T> - struct tuple_size<T const volatile> : public tuple_size<T> {}; - -/*! This metafunction returns the number of elements - * of a \p tuple type of interest. - * - * \tparam T A \c tuple type of interest. - * - * \see pair - * \see tuple - */ -template<class T> - struct tuple_size -{ - /*! The result of this metafunction is returned in \c value. - */ - static const int value = 1 + tuple_size<typename T::tail_type>::value; -}; // end tuple_size - - -// specializations for tuple_size -template<> - struct tuple_size< tuple<> > -{ - static const int value = 0; -}; // end tuple_size< tuple<> > - -template<> - struct tuple_size<null_type> -{ - static const int value = 0; -}; // end tuple_size<null_type> - - - -// forward declaration of detail::cons -namespace detail -{ - -template <class HT, class TT> struct cons; - -} // end detail - - -// -- some traits classes for get functions -template <class T> struct access_traits -{ - typedef const T& const_type; - typedef T& non_const_type; - - typedef const typename thrust::detail::remove_cv<T>::type& parameter_type; - -// used as the tuple constructors parameter types -// Rationale: non-reference tuple element types can be cv-qualified. -// It should be possible to initialize such types with temporaries, -// and when binding temporaries to references, the reference must -// be non-volatile and const. 8.5.3. (5) -}; // end access_traits - -template <class T> struct access_traits<T&> -{ - typedef T& const_type; - typedef T& non_const_type; - - typedef T& parameter_type; -}; // end access_traits<T&> - -// forward declarations of get() -template<int N, class HT, class TT> -THRUST_HOST_DEVICE -inline typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::non_const_type -// XXX we probably don't need to do this for any compiler we care about -jph -//get(cons<HT, TT>& c BOOST_APPEND_EXPLICIT_TEMPLATE_NON_TYPE(int, N)); -get(detail::cons<HT, TT>& c); - -template<int N, class HT, class TT> -THRUST_HOST_DEVICE -inline typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::const_type -// XXX we probably don't need to do this for any compiler we care about -jph -//get(const cons<HT, TT>& c BOOST_APPEND_EXPLICIT_TEMPLATE_NON_TYPE(int, N)); -get(const detail::cons<HT, TT>& c); - -namespace detail -{ - -// -- generate error template, referencing to non-existing members of this -// template is used to produce compilation errors intentionally -template<class T> -class generate_error; - -// - cons getters -------------------------------------------------------- -// called: get_class<N>::get<RETURN_TYPE>(aTuple) - -template< int N > -struct get_class -{ - template<class RET, class HT, class TT > - THRUST_HOST_DEVICE - inline static RET get(const cons<HT, TT>& t) - { - // XXX we may not need to deal with this for any compiler we care about -jph - //return get_class<N-1>::BOOST_NESTED_TEMPLATE get<RET>(t.tail); - return get_class<N-1>::template get<RET>(t.tail); - - // gcc 4.3 couldn't compile this: - //return get_class<N-1>::get<RET>(t.tail); - } - - template<class RET, class HT, class TT > - THRUST_HOST_DEVICE - inline static RET get(cons<HT, TT>& t) - { - // XXX we may not need to deal with this for any compiler we care about -jph - //return get_class<N-1>::BOOST_NESTED_TEMPLATE get<RET>(t.tail); - return get_class<N-1>::template get<RET>(t.tail); - - // gcc 4.3 couldn't compile this: - //return get_class<N-1>::get<RET>(t.tail); - } -}; // end get_class - -template<> -struct get_class<0> -{ - template<class RET, class HT, class TT> - THRUST_HOST_DEVICE - inline static RET get(const cons<HT, TT>& t) - { - return t.head; - } - - template<class RET, class HT, class TT> - THRUST_HOST_DEVICE - inline static RET get(cons<HT, TT>& t) - { - return t.head; - } -}; // get get_class<0> - - -template <bool If, class Then, class Else> struct IF -{ - typedef Then RET; -}; - -template <class Then, class Else> struct IF<false, Then, Else> -{ - typedef Else RET; -}; - -// These helper templates wrap void types and plain function types. -// The rationale is to allow one to write tuple types with those types -// as elements, even though it is not possible to instantiate such object. -// E.g: typedef tuple<void> some_type; // ok -// but: some_type x; // fails - -template <class T> class non_storeable_type -{ - THRUST_HOST_DEVICE - non_storeable_type(); -}; - -template <class T> struct wrap_non_storeable_type -{ - // XXX is_function looks complicated; punt for now -jph - //typedef typename IF< - // ::thrust::detail::is_function<T>::value, non_storeable_type<T>, T - //>::RET type; - - typedef T type; -}; - -template <> struct wrap_non_storeable_type<void> -{ - typedef non_storeable_type<void> type; -}; - - -template <class HT, class TT> - struct cons -{ - typedef HT head_type; - typedef TT tail_type; - - typedef typename - wrap_non_storeable_type<head_type>::type stored_head_type; - - stored_head_type head; - tail_type tail; - - inline THRUST_HOST_DEVICE - typename access_traits<stored_head_type>::non_const_type - get_head() { return head; } - - inline THRUST_HOST_DEVICE - typename access_traits<tail_type>::non_const_type - get_tail() { return tail; } - - inline THRUST_HOST_DEVICE - typename access_traits<stored_head_type>::const_type - get_head() const { return head; } - - inline THRUST_HOST_DEVICE - typename access_traits<tail_type>::const_type - get_tail() const { return tail; } - - inline THRUST_HOST_DEVICE - cons(void) : head(), tail() {} - // cons() : head(detail::default_arg<HT>::f()), tail() {} - - // the argument for head is not strictly needed, but it prevents - // array type elements. This is good, since array type elements - // cannot be supported properly in any case (no assignment, - // copy works only if the tails are exactly the same type, ...) - - inline THRUST_HOST_DEVICE - cons(typename access_traits<stored_head_type>::parameter_type h, - const tail_type& t) - : head (h), tail(t) {} - - template <class T1, class T2, class T3, class T4, class T5, - class T6, class T7, class T8, class T9, class T10> - inline THRUST_HOST_DEVICE - cons( T1& t1, T2& t2, T3& t3, T4& t4, T5& t5, - T6& t6, T7& t7, T8& t8, T9& t9, T10& t10 ) - : head (t1), - tail (t2, t3, t4, t5, t6, t7, t8, t9, t10, static_cast<const null_type&>(null_type())) - {} - - template <class T2, class T3, class T4, class T5, - class T6, class T7, class T8, class T9, class T10> - inline THRUST_HOST_DEVICE - cons( const null_type& /*t1*/, T2& t2, T3& t3, T4& t4, T5& t5, - T6& t6, T7& t7, T8& t8, T9& t9, T10& t10 ) - : head (), - tail (t2, t3, t4, t5, t6, t7, t8, t9, t10, static_cast<const null_type&>(null_type())) - {} - - - template <class HT2, class TT2> - inline THRUST_HOST_DEVICE - cons( const cons<HT2, TT2>& u ) : head(u.head), tail(u.tail) {} - -#if THRUST_CPP_DIALECT >= 2011 - cons(const cons &) = default; -#endif - - __thrust_exec_check_disable__ - template <class HT2, class TT2> - inline THRUST_HOST_DEVICE - cons& operator=( const cons<HT2, TT2>& u ) { - head=u.head; tail=u.tail; return *this; - } - - // must define assignment operator explicitly, implicit version is - // illformed if HT is a reference (12.8. (12)) - __thrust_exec_check_disable__ - inline THRUST_HOST_DEVICE - cons& operator=(const cons& u) { - head = u.head; tail = u.tail; return *this; - } - - // XXX enable when we support std::pair -jph - //template <class T1, class T2> - //THRUST_HOST_DEVICE - //cons& operator=( const std::pair<T1, T2>& u ) { - // //BOOST_STATIC_ASSERT(length<cons>::value == 2); // check length = 2 - // head = u.first; tail.head = u.second; return *this; - //} - - // get member functions (non-const and const) - template <int N> - THRUST_HOST_DEVICE - typename access_traits< - typename tuple_element<N, cons<HT, TT> >::type - >::non_const_type - get() { - return thrust::get<N>(*this); // delegate to non-member get - } - - template <int N> - THRUST_HOST_DEVICE - typename access_traits< - typename tuple_element<N, cons<HT, TT> >::type - >::const_type - get() const { - return thrust::get<N>(*this); // delegate to non-member get - } - - inline THRUST_HOST_DEVICE - void swap(cons &c) - { - using thrust::swap; - - swap(head, c.head); - tail.swap(c.tail); - } -}; - -template <class HT> - struct cons<HT, null_type> -{ - typedef HT head_type; - typedef null_type tail_type; - typedef cons<HT, null_type> self_type; - - typedef typename - wrap_non_storeable_type<head_type>::type stored_head_type; - stored_head_type head; - - typename access_traits<stored_head_type>::non_const_type - inline THRUST_HOST_DEVICE - get_head() { return head; } - - inline THRUST_HOST_DEVICE - null_type get_tail() { return null_type(); } - - inline THRUST_HOST_DEVICE - typename access_traits<stored_head_type>::const_type - get_head() const { return head; } - - inline THRUST_HOST_DEVICE - null_type get_tail() const { return null_type(); } - - inline THRUST_HOST_DEVICE - cons() : head() {} - - inline THRUST_HOST_DEVICE - cons(typename access_traits<stored_head_type>::parameter_type h, - const null_type& = null_type()) - : head (h) {} - - template<class T1> - inline THRUST_HOST_DEVICE - cons(T1& t1, const null_type&, const null_type&, const null_type&, - const null_type&, const null_type&, const null_type&, - const null_type&, const null_type&, const null_type&) - : head (t1) {} - - inline THRUST_HOST_DEVICE - cons(const null_type&, - const null_type&, const null_type&, const null_type&, - const null_type&, const null_type&, const null_type&, - const null_type&, const null_type&, const null_type&) - : head () {} - - template <class HT2> - inline THRUST_HOST_DEVICE - cons( const cons<HT2, null_type>& u ) : head(u.head) {} - -#if THRUST_CPP_DIALECT >= 2011 - cons(const cons &) = default; -#endif - - __thrust_exec_check_disable__ - template <class HT2> - inline THRUST_HOST_DEVICE - cons& operator=(const cons<HT2, null_type>& u ) - { - head = u.head; - return *this; - } - - // must define assignment operator explicitly, implicit version - // is illformed if HT is a reference - inline THRUST_HOST_DEVICE - cons& operator=(const cons& u) { head = u.head; return *this; } - - template <int N> - inline THRUST_HOST_DEVICE - typename access_traits< - typename tuple_element<N, self_type>::type - >::non_const_type - // XXX we probably don't need this for the compilers we care about -jph - //get(BOOST_EXPLICIT_TEMPLATE_NON_TYPE(int, N)) - get(void) - { - return thrust::get<N>(*this); - } - - template <int N> - inline THRUST_HOST_DEVICE - typename access_traits< - typename tuple_element<N, self_type>::type - >::const_type - // XXX we probably don't need this for the compilers we care about -jph - //get(BOOST_EXPLICIT_TEMPLATE_NON_TYPE(int, N)) const - get(void) const - { - return thrust::get<N>(*this); - } - - inline THRUST_HOST_DEVICE - void swap(cons &c) - { - using thrust::swap; - - swap(head, c.head); - } -}; // end cons - -template <class T0, class T1, class T2, class T3, class T4, - class T5, class T6, class T7, class T8, class T9> - struct map_tuple_to_cons -{ - typedef cons<T0, - typename map_tuple_to_cons<T1, T2, T3, T4, T5, - T6, T7, T8, T9, null_type>::type - > type; -}; // end map_tuple_to_cons - -// The empty tuple is a null_type -template <> - struct map_tuple_to_cons<null_type, null_type, null_type, null_type, null_type, null_type, null_type, null_type, null_type, null_type> -{ - typedef null_type type; -}; // end map_tuple_to_cons<...> - - - -// --------------------------------------------------------------------------- -// The call_traits for make_tuple - -// Must be instantiated with plain or const plain types (not with references) - -// from template<class T> foo(const T& t) : make_tuple_traits<const T>::type -// from template<class T> foo(T& t) : make_tuple_traits<T>::type - -// Conversions: -// T -> T, -// references -> compile_time_error -// array -> const ref array - - -template<class T> -struct make_tuple_traits { - typedef T type; - - // commented away, see below (JJ) - // typedef typename IF< - // boost::is_function<T>::value, - // T&, - // T>::RET type; - -}; - -// The is_function test was there originally for plain function types, -// which can't be stored as such (we must either store them as references or -// pointers). Such a type could be formed if make_tuple was called with a -// reference to a function. -// But this would mean that a const qualified function type was formed in -// the make_tuple function and hence make_tuple can't take a function -// reference as a parameter, and thus T can't be a function type. -// So is_function test was removed. -// (14.8.3. says that type deduction fails if a cv-qualified function type -// is created. (It only applies for the case of explicitly specifying template -// args, though?)) (JJ) - -template<class T> -struct make_tuple_traits<T&> { - typedef typename - detail::generate_error<T&>:: - do_not_use_with_reference_type error; -}; - -// Arrays can't be stored as plain types; convert them to references. -// All arrays are converted to const. This is because make_tuple takes its -// parameters as const T& and thus the knowledge of the potential -// non-constness of actual argument is lost. -template<class T, int n> struct make_tuple_traits <T[n]> { - typedef const T (&type)[n]; -}; - -template<class T, int n> -struct make_tuple_traits<const T[n]> { - typedef const T (&type)[n]; -}; - -template<class T, int n> struct make_tuple_traits<volatile T[n]> { - typedef const volatile T (&type)[n]; -}; - -template<class T, int n> -struct make_tuple_traits<const volatile T[n]> { - typedef const volatile T (&type)[n]; -}; - -// XXX enable these if we ever care about reference_wrapper -jph -//template<class T> -//struct make_tuple_traits<reference_wrapper<T> >{ -// typedef T& type; -//}; -// -//template<class T> -//struct make_tuple_traits<const reference_wrapper<T> >{ -// typedef T& type; -//}; - - -// a helper traits to make the make_tuple functions shorter (Vesa Karvonen's -// suggestion) -template < - class T0 = null_type, class T1 = null_type, class T2 = null_type, - class T3 = null_type, class T4 = null_type, class T5 = null_type, - class T6 = null_type, class T7 = null_type, class T8 = null_type, - class T9 = null_type -> -struct make_tuple_mapper { - typedef - tuple<typename make_tuple_traits<T0>::type, - typename make_tuple_traits<T1>::type, - typename make_tuple_traits<T2>::type, - typename make_tuple_traits<T3>::type, - typename make_tuple_traits<T4>::type, - typename make_tuple_traits<T5>::type, - typename make_tuple_traits<T6>::type, - typename make_tuple_traits<T7>::type, - typename make_tuple_traits<T8>::type, - typename make_tuple_traits<T9>::type> type; -}; - -} // end detail - - -template<int N, class HT, class TT> -THRUST_HOST_DEVICE -inline typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::non_const_type -get(detail::cons<HT, TT>& c) -{ - //return detail::get_class<N>::BOOST_NESTED_TEMPLATE - - // gcc 4.3 couldn't compile this: - //return detail::get_class<N>:: - - return detail::get_class<N>::template - get< - typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::non_const_type, - HT,TT - >(c); -} - - -// get function for const cons-lists, returns a const reference to -// the element. If the element is a reference, returns the reference -// as such (that is, can return a non-const reference) -template<int N, class HT, class TT> -THRUST_HOST_DEVICE -inline typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::const_type -get(const detail::cons<HT, TT>& c) -{ - //return detail::get_class<N>::BOOST_NESTED_TEMPLATE - - // gcc 4.3 couldn't compile this: - //return detail::get_class<N>:: - - return detail::get_class<N>::template - get< - typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::const_type, - HT,TT - >(c); -} - - -template<class T0> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0>::type - make_tuple(const T0& t0) -{ - typedef typename detail::make_tuple_mapper<T0>::type t; - return t(t0); -} // end make_tuple() - -template<class T0, class T1> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1>::type - make_tuple(const T0& t0, const T1& t1) -{ - typedef typename detail::make_tuple_mapper<T0,T1>::type t; - return t(t0,t1); -} // end make_tuple() - -template<class T0, class T1, class T2> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2>::type t; - return t(t0,t1,t2); -} // end make_tuple() - -template<class T0, class T1, class T2, class T3> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2,T3>::type t; - return t(t0,t1,t2,t3); -} // end make_tuple() - -template<class T0, class T1, class T2, class T3, class T4> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2,T3,T4>::type t; - return t(t0,t1,t2,t3,t4); -} // end make_tuple() - -template<class T0, class T1, class T2, class T3, class T4, class T5> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2,T3,T4,T5>::type t; - return t(t0,t1,t2,t3,t4,t5); -} // end make_tuple() - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2,T3,T4,T5,T6>::type t; - return t(t0,t1,t2,t3,t4,t5,t6); -} // end make_tuple() - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6, class T7> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6, T7>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6, const T7& t7) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2,T3,T4,T5,T6,T7>::type t; - return t(t0,t1,t2,t3,t4,t5,t6,t7); -} // end make_tuple() - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6, class T7, class T8> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6, T7, T8>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6, const T7& t7, const T8& t8) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2,T3,T4,T5,T6,T7,T8>::type t; - return t(t0,t1,t2,t3,t4,t5,t6,t7,t8); -} // end make_tuple() - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6, class T7, class T8, class T9> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6, T7, T8, T9>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6, const T7& t7, const T8& t8, const T9& t9) -{ - typedef typename detail::make_tuple_mapper<T0,T1,T2,T3,T4,T5,T6,T7,T8,T9>::type t; - return t(t0,t1,t2,t3,t4,t5,t6,t7,t8,t9); -} // end make_tuple() - - -template<typename T0> -THRUST_HOST_DEVICE inline -tuple<T0&> tie(T0 &t0) -{ - return tuple<T0&>(t0); -} - -template<typename T0,typename T1> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&> tie(T0 &t0, T1 &t1) -{ - return tuple<T0&,T1&>(t0,t1); -} - -template<typename T0,typename T1, typename T2> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&> tie(T0 &t0, T1 &t1, T2 &t2) -{ - return tuple<T0&,T1&,T2&>(t0,t1,t2); -} - -template<typename T0,typename T1, typename T2, typename T3> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3) -{ - return tuple<T0&,T1&,T2&,T3&>(t0,t1,t2,t3); -} - -template<typename T0,typename T1, typename T2, typename T3, typename T4> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4) -{ - return tuple<T0&,T1&,T2&,T3&,T4&>(t0,t1,t2,t3,t4); -} - -template<typename T0,typename T1, typename T2, typename T3, typename T4, typename T5> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5) -{ - return tuple<T0&,T1&,T2&,T3&,T4&,T5&>(t0,t1,t2,t3,t4,t5); -} - -template<typename T0,typename T1, typename T2, typename T3, typename T4, typename T5, typename T6> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6) -{ - return tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&>(t0,t1,t2,t3,t4,t5,t6); -} - -template<typename T0,typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6, T7 &t7) -{ - return tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&>(t0,t1,t2,t3,t4,t5,t6,t7); -} - -template<typename T0,typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&,T8&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6, T7 &t7, T8 &t8) -{ - return tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&,T8&>(t0,t1,t2,t3,t4,t5,t6,t7,t8); -} - -template<typename T0,typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8, typename T9> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&,T8&,T9&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6, T7 &t7, T8 &t8, T9 &t9) -{ - return tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&,T8&,T9&>(t0,t1,t2,t3,t4,t5,t6,t7,t8,t9); -} - -template< - typename T0, typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8, typename T9, - typename U0, typename U1, typename U2, typename U3, typename U4, typename U5, typename U6, typename U7, typename U8, typename U9 -> -THRUST_HOST_DEVICE inline -void swap(thrust::tuple<T0,T1,T2,T3,T4,T5,T6,T7,T8,T9> &x, - thrust::tuple<U0,U1,U2,U3,U4,U5,U6,U7,U8,U9> &y) -{ - return x.swap(y); -} - - - -namespace detail -{ - -template<class T1, class T2> -THRUST_HOST_DEVICE -inline bool eq(const T1& lhs, const T2& rhs) { - return lhs.get_head() == rhs.get_head() && - eq(lhs.get_tail(), rhs.get_tail()); -} -template<> -THRUST_HOST_DEVICE -inline bool eq<null_type,null_type>(const null_type&, const null_type&) { return true; } - -template<class T1, class T2> -THRUST_HOST_DEVICE -inline bool neq(const T1& lhs, const T2& rhs) { - return lhs.get_head() != rhs.get_head() || - neq(lhs.get_tail(), rhs.get_tail()); -} -template<> -THRUST_HOST_DEVICE -inline bool neq<null_type,null_type>(const null_type&, const null_type&) { return false; } - -template<class T1, class T2> -THRUST_HOST_DEVICE -inline bool lt(const T1& lhs, const T2& rhs) { - return (lhs.get_head() < rhs.get_head()) || - (!(rhs.get_head() < lhs.get_head()) && - lt(lhs.get_tail(), rhs.get_tail())); -} -template<> -THRUST_HOST_DEVICE -inline bool lt<null_type,null_type>(const null_type&, const null_type&) { return false; } - -template<class T1, class T2> -THRUST_HOST_DEVICE -inline bool gt(const T1& lhs, const T2& rhs) { - return (lhs.get_head() > rhs.get_head()) || - (!(rhs.get_head() > lhs.get_head()) && - gt(lhs.get_tail(), rhs.get_tail())); -} -template<> -THRUST_HOST_DEVICE -inline bool gt<null_type,null_type>(const null_type&, const null_type&) { return false; } - -template<class T1, class T2> -THRUST_HOST_DEVICE -inline bool lte(const T1& lhs, const T2& rhs) { - return lhs.get_head() <= rhs.get_head() && - ( !(rhs.get_head() <= lhs.get_head()) || - lte(lhs.get_tail(), rhs.get_tail())); -} -template<> -THRUST_HOST_DEVICE -inline bool lte<null_type,null_type>(const null_type&, const null_type&) { return true; } - -template<class T1, class T2> -THRUST_HOST_DEVICE -inline bool gte(const T1& lhs, const T2& rhs) { - return lhs.get_head() >= rhs.get_head() && - ( !(rhs.get_head() >= lhs.get_head()) || - gte(lhs.get_tail(), rhs.get_tail())); -} -template<> -THRUST_HOST_DEVICE -inline bool gte<null_type,null_type>(const null_type&, const null_type&) { return true; } - -} // end detail - - - -// equal ---- - -template<class T1, class T2, class S1, class S2> -THRUST_HOST_DEVICE -inline bool operator==(const detail::cons<T1, T2>& lhs, const detail::cons<S1, S2>& rhs) -{ - // XXX support this eventually -jph - //// check that tuple lengths are equal - //BOOST_STATIC_ASSERT(tuple_size<T2>::value == tuple_size<S2>::value); - - return detail::eq(lhs, rhs); -} // end operator==() - -// not equal ----- - -template<class T1, class T2, class S1, class S2> -THRUST_HOST_DEVICE -inline bool operator!=(const detail::cons<T1, T2>& lhs, const detail::cons<S1, S2>& rhs) -{ - // XXX support this eventually -jph - //// check that tuple lengths are equal - //BOOST_STATIC_ASSERT(tuple_size<T2>::value == tuple_size<S2>::value); - - return detail::neq(lhs, rhs); -} // end operator!=() - -// < -template<class T1, class T2, class S1, class S2> -THRUST_HOST_DEVICE -inline bool operator<(const detail::cons<T1, T2>& lhs, const detail::cons<S1, S2>& rhs) -{ - // XXX support this eventually -jph - //// check that tuple lengths are equal - //BOOST_STATIC_ASSERT(tuple_size<T2>::value == tuple_size<S2>::value); - - return detail::lt(lhs, rhs); -} // end operator<() - -// > -template<class T1, class T2, class S1, class S2> -THRUST_HOST_DEVICE -inline bool operator>(const detail::cons<T1, T2>& lhs, const detail::cons<S1, S2>& rhs) -{ - // XXX support this eventually -jph - //// check that tuple lengths are equal - //BOOST_STATIC_ASSERT(tuple_size<T2>::value == tuple_size<S2>::value); - - return detail::gt(lhs, rhs); -} // end operator>() - -// <= -template<class T1, class T2, class S1, class S2> -THRUST_HOST_DEVICE -inline bool operator<=(const detail::cons<T1, T2>& lhs, const detail::cons<S1, S2>& rhs) -{ - // XXX support this eventually -jph - //// check that tuple lengths are equal - //BOOST_STATIC_ASSERT(tuple_size<T2>::value == tuple_size<S2>::value); - - return detail::lte(lhs, rhs); -} // end operator<=() - -// >= -template<class T1, class T2, class S1, class S2> -THRUST_HOST_DEVICE -inline bool operator>=(const detail::cons<T1, T2>& lhs, const detail::cons<S1, S2>& rhs) -{ - // XXX support this eventually -jph - //// check that tuple lengths are equal - //BOOST_STATIC_ASSERT(tuple_size<T2>::value == tuple_size<S2>::value); - - return detail::gte(lhs, rhs); -} // end operator>=() - -THRUST_NAMESPACE_END - diff --git a/thrust/iterator/detail/tuple_of_iterator_references.h b/thrust/iterator/detail/tuple_of_iterator_references.h index 26f99740..d8e18c3b 100644 --- a/thrust/iterator/detail/tuple_of_iterator_references.h +++ b/thrust/iterator/detail/tuple_of_iterator_references.h @@ -17,130 +17,100 @@ #pragma once #include <thrust/detail/config.h> -#include <thrust/tuple.h> -#include <thrust/pair.h> #include <thrust/detail/reference_forward_declaration.h> +#include <thrust/pair.h> +#include <thrust/tuple.h> THRUST_NAMESPACE_BEGIN namespace detail { - -template< - typename... Ts -> - class tuple_of_iterator_references - : public thrust::tuple<Ts...> +template <typename... Ts> +class tuple_of_iterator_references : public thrust::tuple<Ts...> { - private: - typedef thrust::tuple<Ts...> super_t; + using super_t = thrust::tuple<Ts...>; + using super_t::super_t; - public: - // allow implicit construction from tuple<refs> - inline THRUST_HOST_DEVICE - tuple_of_iterator_references(const super_t &other) +public: + // allow implicit construction from tuple<refs> + inline THRUST_HOST_DEVICE + tuple_of_iterator_references(const super_t& other) : super_t(other) - {} - - // allow assignment from tuples - // XXX might be worthwhile to guard this with an enable_if is_assignable - __thrust_exec_check_disable__ - template<typename... Us> - inline THRUST_HOST_DEVICE - tuple_of_iterator_references &operator=(const thrust::tuple<Us...> &other) - { - super_t::operator=(other); - return *this; - } - - // allow assignment from pairs - // XXX might be worthwhile to guard this with an enable_if is_assignable - __thrust_exec_check_disable__ - template<typename U1, typename U2> - inline THRUST_HOST_DEVICE - tuple_of_iterator_references &operator=(const thrust::pair<U1,U2> &other) - { - super_t::operator=(other); - return *this; - } - - // allow assignment from reference<tuple> - // XXX perhaps we should generalize to reference<T> - // we could captures reference<pair> this way - __thrust_exec_check_disable__ - template<typename Pointer, typename Derived, - typename... Us> - inline THRUST_HOST_DEVICE -// XXX gcc-4.2 crashes on is_assignable -// typename thrust::detail::enable_if< -// thrust::detail::is_assignable< -// super_t, -// const thrust::tuple<Us...> -// >::value, -// tuple_of_iterator_references & -// >::type - tuple_of_iterator_references & - operator=(const thrust::reference<thrust::tuple<Us...>, Pointer, Derived> &other) - { - typedef thrust::tuple<Us...> tuple_type; - - // XXX perhaps this could be accelerated - tuple_type other_tuple = other; - super_t::operator=(other_tuple); - return *this; - } - - - // duplicate thrust::tuple's constructors - inline THRUST_HOST_DEVICE - tuple_of_iterator_references() {} - - inline THRUST_HOST_DEVICE - tuple_of_iterator_references(typename access_traits<Ts>::parameter_type... ts) - : super_t(ts...) - {} + {} + + // allow assignment from tuples + __thrust_exec_check_disable__ + template <typename... Us> + inline THRUST_HOST_DEVICE + tuple_of_iterator_references& operator=(const thrust::tuple<Us...>& other) + { + super_t::operator=(other); + return *this; + } + + // allow assignment from pairs + __thrust_exec_check_disable__ + template <typename U1, typename U2> + inline THRUST_HOST_DEVICE + tuple_of_iterator_references& operator=(const thrust::pair<U1, U2>& other) + { + get<0>(*this) = other.first; + get<1>(*this) = other.second; + return *this; + } + + // allow assignment from reference<tuple> + // XXX perhaps we should generalize to reference<T> + // we could captures reference<pair> this way + __thrust_exec_check_disable__ + template <typename Pointer, typename Derived, typename... Us> + inline THRUST_HOST_DEVICE + tuple_of_iterator_references& + operator=(const thrust::reference<thrust::tuple<Us...>, Pointer, Derived>& other) + { + typedef thrust::tuple<Us...> tuple_type; + + // XXX perhaps this could be accelerated + tuple_type other_tuple = other; + super_t::operator=(other_tuple); + return *this; + } + + // allow conversion to tuple + // XXX perhaps we should constraint with enable_if + template <class... Us> + inline THRUST_HOST_DEVICE + operator thrust::tuple<Us...>() const { return thrust::tuple<Us...>{*this}; } }; +} // namespace detail -// this overload of swap() permits swapping tuple_of_iterator_references returned as temporaries from -// iterator dereferences -template< - typename... Ts, - typename... Us -> -inline THRUST_HOST_DEVICE -void swap(tuple_of_iterator_references<Ts...> x, - tuple_of_iterator_references<Us...> y) -{ - x.swap(y); -} - +THRUST_NAMESPACE_END -} // end detail +_LIBCUDACXX_BEGIN_NAMESPACE_STD // define tuple_size, tuple_element, etc. -template<class... Ts> -struct tuple_size<detail::tuple_of_iterator_references<Ts...>> - : std::integral_constant<size_t, sizeof...(Ts)> +template <class... Ts> +struct tuple_size<THRUST_NS_QUALIFIER::detail::tuple_of_iterator_references<Ts...>> + : std::integral_constant<size_t, sizeof...(Ts)> {}; -template<size_t i> -struct tuple_element<i, detail::tuple_of_iterator_references<>> {}; - +template <size_t i> +struct tuple_element<i, THRUST_NS_QUALIFIER::detail::tuple_of_iterator_references<>> +{}; -template<class T, class... Ts> -struct tuple_element<0, detail::tuple_of_iterator_references<T,Ts...>> +template <class T, class... Ts> +struct tuple_element<0, THRUST_NS_QUALIFIER::detail::tuple_of_iterator_references<T, Ts...>> { using type = T; }; - -template<size_t i, class T, class... Ts> -struct tuple_element<i, detail::tuple_of_iterator_references<T,Ts...>> +template <size_t i, class T, class... Ts> +struct tuple_element<i, THRUST_NS_QUALIFIER::detail::tuple_of_iterator_references<T, Ts...>> { - using type = typename tuple_element<i - 1, detail::tuple_of_iterator_references<Ts...>>::type; + using type = + typename tuple_element<i - 1, + THRUST_NS_QUALIFIER::detail::tuple_of_iterator_references<Ts...>>::type; }; - -THRUST_NAMESPACE_END - +_LIBCUDACXX_END_NAMESPACE_STD diff --git a/thrust/pair.h b/thrust/pair.h index cbce9c57..c7fe10a8 100644 --- a/thrust/pair.h +++ b/thrust/pair.h @@ -21,7 +21,14 @@ #pragma once #include <thrust/detail/config.h> -#include <utility> + +#include <cuda/std/utility> + +namespace std +{ +template <class T, class U> +class pair; +} THRUST_NAMESPACE_BEGIN @@ -33,6 +40,25 @@ THRUST_NAMESPACE_BEGIN * \{ */ +/*! This convenience metafunction is included for compatibility with + * \p tuple. It returns either the type of a \p pair's + * \c first_type or \c second_type in its nested type, \c type. + * + * \tparam N This parameter selects the member of interest. + * \tparam T A \c pair type of interest. + */ +template <size_t N, class T> +using tuple_element = ::cuda::std::tuple_element<N, T>; + +/*! This convenience metafunction is included for compatibility with + * \p tuple. It returns \c 2, the number of elements of a \p pair, + * in its nested data member, \c value. + * + * \tparam Pair A \c pair type of interest. + */ +template <class T> +using tuple_size = ::cuda::std::tuple_size<T>; + /*! \p pair is a generic data structure encapsulating a heterogeneous * pair of values. * @@ -44,167 +70,81 @@ THRUST_NAMESPACE_BEGIN * requirements on the type of \p T2. <tt>T2</tt>'s type is * provided by <tt>pair::second_type</tt>. */ -template <typename T1, typename T2> - struct pair +template <class T, class U> +class pair : public ::cuda::std::pair<T, U> { - /*! \p first_type is the type of \p pair's first object type. - */ - typedef T1 first_type; - - /*! \p second_type is the type of \p pair's second object type. - */ - typedef T2 second_type; +public: + using super_t = ::cuda::std::pair<T, U>; + using super_t::super_t; - /*! The \p pair's first object. - */ - first_type first; - - /*! The \p pair's second object. - */ - second_type second; - - /*! \p pair's default constructor constructs \p first - * and \p second using \c first_type & \c second_type's - * default constructors, respectively. - */ - THRUST_HOST_DEVICE pair(void); - - /*! This constructor accepts two objects to copy into this \p pair. - * - * \param x The object to copy into \p first. - * \param y The object to copy into \p second. - */ + // allow construction from std::pair + template <class T2, class U2> inline THRUST_HOST_DEVICE - pair(const T1 &x, const T2 &y); + pair(const ::std::pair<T2, U2>& other) + : super_t(other.first, other.second) + {} - /*! This copy constructor copies from a \p pair whose types are - * convertible to this \p pair's \c first_type and \c second_type, - * respectively. - * - * \param p The \p pair to copy from. - * - * \tparam U1 is convertible to \c first_type. - * \tparam U2 is convertible to \c second_type. - */ - template <typename U1, typename U2> - inline THRUST_HOST_DEVICE - pair(const pair<U1,U2> &p); - - /*! This copy constructor copies from a <tt>std::pair</tt> whose types are - * convertible to this \p pair's \c first_type and \c second_type, - * respectively. - * - * \param p The <tt>std::pair</tt> to copy from. - * - * \tparam U1 is convertible to \c first_type. - * \tparam U2 is convertible to \c second_type. - */ - template <typename U1, typename U2> - inline THRUST_HOST_DEVICE - pair(const std::pair<U1,U2> &p); - - /*! \p swap swaps the elements of two <tt>pair</tt>s. - * - * \param p The other <tt>pair</tt> with which to swap. - */ - inline THRUST_HOST_DEVICE - void swap(pair &p); -}; // end pair - - -/*! This operator tests two \p pairs for equality. - * - * \param x The first \p pair to compare. - * \param y The second \p pair to compare. - * \return \c true if and only if <tt>x.first == y.first && x.second == y.second</tt>. - * - * \tparam T1 is a model of <a href="https://en.cppreference.com/w/cpp/concepts/equality_comparable">Equality Comparable</a>. - * \tparam T2 is a model of <a href="https://en.cppreference.com/w/cpp/concepts/equality_comparable">Equality Comparable</a>. - */ -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator==(const pair<T1,T2> &x, const pair<T1,T2> &y); - - -/*! This operator tests two pairs for ascending ordering. - * - * \param x The first \p pair to compare. - * \param y The second \p pair to compare. - * \return \c true if and only if <tt>x.first < y.first || (!(y.first < x.first) && x.second < y.second)</tt>. - * - * \tparam T1 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - * \tparam T2 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - */ -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator<(const pair<T1,T2> &x, const pair<T1,T2> &y); - - -/*! This operator tests two pairs for inequality. - * - * \param x The first \p pair to compare. - * \param y The second \p pair to compare. - * \return \c true if and only if <tt>!(x == y)</tt>. - * - * \tparam T1 is a model of <a href="https://en.cppreference.com/w/cpp/concepts/equality_comparable">Equality Comparable</a>. - * \tparam T2 is a model of <a href="https://en.cppreference.com/w/cpp/concepts/equality_comparable">Equality Comparable</a>. - */ -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator!=(const pair<T1,T2> &x, const pair<T1,T2> &y); - - -/*! This operator tests two pairs for descending ordering. - * - * \param x The first \p pair to compare. - * \param y The second \p pair to compare. - * \return \c true if and only if <tt>y < x</tt>. - * - * \tparam T1 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - * \tparam T2 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - */ -template <typename T1, typename T2> + // allow assignment from std::pair + template <class T2, class U2> inline THRUST_HOST_DEVICE - bool operator>(const pair<T1,T2> &x, const pair<T1,T2> &y); - + pair& operator=(const ::std::pair<T2, U2>& other) + { + this->first = other.first; + this->second = other.second; + return *this; + } +}; + +// We cannot derive from `cuda::std::pair` directly, so we need to specialize `get` +template <size_t N, class T, class U> +inline THRUST_HOST_DEVICE +typename tuple_element<N, pair<T, U>>::type& get(pair<T, U>& t) noexcept +{ + using ::cuda::std::get; + return get<N>(t); +} -/*! This operator tests two pairs for ascending ordering or equivalence. - * - * \param x The first \p pair to compare. - * \param y The second \p pair to compare. - * \return \c true if and only if <tt>!(y < x)</tt>. - * - * \tparam T1 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - * \tparam T2 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - */ -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator<=(const pair<T1,T2> &x, const pair<T1,T2> &y); +template <size_t N, class T, class U> +inline THRUST_HOST_DEVICE +typename tuple_element<N, pair<T, U>>::type&& get(pair<T, U>&& t) noexcept +{ + using ::cuda::std::get; + return get<N>(static_cast<pair<T, U>&&>(t)); +} +template <size_t N, class T, class U> +inline THRUST_HOST_DEVICE +const typename tuple_element<N, pair<T, U>>::type& get(const pair<T, U>& t) noexcept +{ + using ::cuda::std::get; + return get<N>(t); +} -/*! This operator tests two pairs for descending ordering or equivalence. - * - * \param x The first \p pair to compare. - * \param y The second \p pair to compare. - * \return \c true if and only if <tt>!(x < y)</tt>. - * - * \tparam T1 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - * \tparam T2 is a model of <a href="https://en.cppreference.com/w/cpp/named_req/LessThanComparable">LessThan Comparable</a>. - */ -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - bool operator>=(const pair<T1,T2> &x, const pair<T1,T2> &y); +template <size_t N, class T, class U> +inline THRUST_HOST_DEVICE +const typename tuple_element<N, pair<T, U>>::type&& get(const pair<T, U>&& t) noexcept +{ + using ::cuda::std::get; + return get<N>(static_cast<const pair<T, U>&&>(t)); +} +namespace detail +{ +template <class T> +struct unwrap_refwrapper +{ + using type = T; +}; -/*! \p swap swaps the contents of two <tt>pair</tt>s. - * - * \param x The first \p pair to swap. - * \param y The second \p pair to swap. - */ -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - void swap(pair<T1,T2> &x, pair<T1,T2> &y); +template <class T> +struct unwrap_refwrapper<std::reference_wrapper<T>> +{ + using type = T&; +}; +template <class T> +using unwrap_decay_t = typename unwrap_refwrapper<typename ::cuda::std::decay<T>::type>::type; +} // namespace detail /*! This convenience function creates a \p pair from two objects. * @@ -215,67 +155,41 @@ template <typename T1, typename T2> * \tparam T1 There are no requirements on the type of \p T1. * \tparam T2 There are no requirements on the type of \p T2. */ -template <typename T1, typename T2> - inline THRUST_HOST_DEVICE - pair<T1,T2> make_pair(T1 x, T2 y); - +template <class T, class U> +inline THRUST_HOST_DEVICE +pair<detail::unwrap_decay_t<T>, detail::unwrap_decay_t<U>> make_pair(T&& first, U&& second) +{ + return {::cuda::std::forward<T>(first), ::cuda::std::forward<U>(second)}; +} -/*! This convenience metafunction is included for compatibility with - * \p tuple. It returns either the type of a \p pair's - * \c first_type or \c second_type in its nested type, \c type. - * - * \tparam N This parameter selects the member of interest. - * \tparam T A \c pair type of interest. +/*! \endcond */ -template<size_t N, class T> struct tuple_element; - -/*! This convenience metafunction is included for compatibility with - * \p tuple. It returns \c 2, the number of elements of a \p pair, - * in its nested data member, \c value. - * - * \tparam Pair A \c pair type of interest. +/*! \} // tuple */ -template<typename Pair> struct tuple_size; - -/*! This convenience function returns a reference to either the first or - * second member of a \p pair. - * - * \param p The \p pair of interest. - * \return \c p.first or \c p.second, depending on the template - * parameter. - * - * \tparam N This parameter selects the member of interest. +/*! \} // utility */ -// XXX comment out these prototypes as a WAR to a problem on MSVC 2005 -//template<unsigned int N, typename T1, typename T2> -// inline THRUST_HOST_DEVICE -// typename tuple_element<N, pair<T1,T2> >::type & -// get(pair<T1,T2> &p); +THRUST_NAMESPACE_END -/*! This convenience function returns a const reference to either the - * first or second member of a \p pair. - * - * \param p The \p pair of interest. - * \return \c p.first or \c p.second, depending on the template - * parameter. - * - * \tparam i This parameter selects the member of interest. - */ -// XXX comment out these prototypes as a WAR to a problem on MSVC 2005 -//template<int N, typename T1, typename T2> -// inline THRUST_HOST_DEVICE -// const typename tuple_element<N, pair<T1,T2> >::type & -// get(const pair<T1,T2> &p); +_LIBCUDACXX_BEGIN_NAMESPACE_STD -/*! \} // pair - */ +// define tuple_size, tuple_element, etc. +template <class T, class U> +struct tuple_size<THRUST_NS_QUALIFIER::pair<T, U>> : std::integral_constant<size_t, 2> +{}; -/*! \} // utility - */ +template <class T, class U> +struct tuple_element<0, THRUST_NS_QUALIFIER::pair<T, U>> +{ + using type = T; +}; -THRUST_NAMESPACE_END +template <class T, class U> +struct tuple_element<1, THRUST_NS_QUALIFIER::pair<T, U>> +{ + using type = U; +}; -#include <thrust/detail/pair.inl> +_LIBCUDACXX_END_NAMESPACE_STD diff --git a/thrust/tuple.h b/thrust/tuple.h index 8750734a..1613ca27 100644 --- a/thrust/tuple.h +++ b/thrust/tuple.h @@ -14,7 +14,6 @@ * limitations under the License. */ - /*! \file tuple.h * \brief A type encapsulating a heterogeneous collection of elements. */ @@ -31,8 +30,9 @@ #pragma once #include <thrust/detail/config.h> -#include <thrust/detail/tuple.inl> -#include <thrust/pair.h> + +#include <cuda/std/tuple> +#include <cuda/std/utility> THRUST_NAMESPACE_BEGIN @@ -44,14 +44,6 @@ THRUST_NAMESPACE_BEGIN * \{ */ -/*! \cond - */ - -struct null_type; - -/*! \endcond - */ - /*! This metafunction returns the type of a * \p tuple's <tt>N</tt>th element. * @@ -61,7 +53,8 @@ struct null_type; * \see pair * \see tuple */ -template <size_t N, class T> struct tuple_element; +template <size_t N, class T> +using tuple_element = ::cuda::std::tuple_element<N, T>; /*! This metafunction returns the number of elements * of a \p tuple type of interest. @@ -71,73 +64,8 @@ template <size_t N, class T> struct tuple_element; * \see pair * \see tuple */ -template <class T> struct tuple_size; - - -// get function for non-const cons-lists, returns a reference to the element - -/*! The \p get function returns a reference to a \p tuple element of - * interest. - * - * \param t A reference to a \p tuple of interest. - * \return A reference to \p t's <tt>N</tt>th element. - * - * \tparam N The index of the element of interest. - * - * The following code snippet demonstrates how to use \p get to print - * the value of a \p tuple element. - * - * \code - * #include <thrust/tuple.h> - * #include <iostream> - * ... - * thrust::tuple<int, const char *> t(13, "thrust"); - * - * std::cout << "The 1st value of t is " << thrust::get<0>(t) << std::endl; - * \endcode - * - * \see pair - * \see tuple - */ -template<int N, class HT, class TT> -THRUST_HOST_DEVICE -inline typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::non_const_type -get(detail::cons<HT, TT>& t); - - -/*! The \p get function returns a \c const reference to a \p tuple element of - * interest. - * - * \param t A reference to a \p tuple of interest. - * \return A \c const reference to \p t's <tt>N</tt>th element. - * - * \tparam N The index of the element of interest. - * - * The following code snippet demonstrates how to use \p get to print - * the value of a \p tuple element. - * - * \code - * #include <thrust/tuple.h> - * #include <iostream> - * ... - * thrust::tuple<int, const char *> t(13, "thrust"); - * - * std::cout << "The 1st value of t is " << thrust::get<0>(t) << std::endl; - * \endcode - * - * \see pair - * \see tuple - */ -template<int N, class HT, class TT> -THRUST_HOST_DEVICE -inline typename access_traits< - typename tuple_element<N, detail::cons<HT, TT> >::type - >::const_type -get(const detail::cons<HT, TT>& t); - - +template <class T> +using tuple_size = ::cuda::std::tuple_size<T>; /*! \brief \p tuple is a class template that can be instantiated with up to ten * arguments. Each template argument specifies the type of element in the \p @@ -155,7 +83,7 @@ get(const detail::cons<HT, TT>& t); * \code * #include <thrust/tuple.h> * #include <iostream> - * + * * int main() { * // Create a tuple containing an `int`, a `float`, and a string. * thrust::tuple<int, float, const char*> t(13, 0.1f, "thrust"); @@ -178,390 +106,12 @@ get(const detail::cons<HT, TT>& t); * \see tuple_size * \see tie */ -template <class T0, class T1, class T2, class T3, class T4, - class T5, class T6, class T7, class T8, class T9> - class tuple - /*! \cond - */ - : public detail::map_tuple_to_cons<T0, T1, T2, T3, T4, T5, T6, T7, T8, T9>::type - /*! \endcond - */ -{ - /*! \cond - */ - - private: - typedef typename detail::map_tuple_to_cons<T0, T1, T2, T3, T4, T5, T6, T7, T8, T9>::type inherited; - - /*! \endcond - */ - - public: - - /*! \p tuple's no-argument constructor initializes each element. - */ - inline THRUST_HOST_DEVICE - tuple(void) {} - - /*! \p tuple's one-argument constructor copy constructs the first element from the given parameter - * and intializes all other elements. - * \param t0 The value to assign to this \p tuple's first element. - */ - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0) - : inherited(t0, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - /*! \p tuple's one-argument constructor copy constructs the first two elements from the given parameters - * and intializes all other elements. - * \param t0 The value to assign to this \p tuple's first element. - * \param t1 The value to assign to this \p tuple's second element. - * \note \p tuple's constructor has ten variants of this form, the rest of which are ommitted here for brevity. - */ - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1) - : inherited(t0, t1, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - /*! \cond - */ - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2) - : inherited(t0, t1, t2, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2, - typename access_traits<T3>::parameter_type t3) - : inherited(t0, t1, t2, t3, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2, - typename access_traits<T3>::parameter_type t3, - typename access_traits<T4>::parameter_type t4) - : inherited(t0, t1, t2, t3, t4, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2, - typename access_traits<T3>::parameter_type t3, - typename access_traits<T4>::parameter_type t4, - typename access_traits<T5>::parameter_type t5) - : inherited(t0, t1, t2, t3, t4, t5, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2, - typename access_traits<T3>::parameter_type t3, - typename access_traits<T4>::parameter_type t4, - typename access_traits<T5>::parameter_type t5, - typename access_traits<T6>::parameter_type t6) - : inherited(t0, t1, t2, t3, t4, t5, t6, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2, - typename access_traits<T3>::parameter_type t3, - typename access_traits<T4>::parameter_type t4, - typename access_traits<T5>::parameter_type t5, - typename access_traits<T6>::parameter_type t6, - typename access_traits<T7>::parameter_type t7) - : inherited(t0, t1, t2, t3, t4, t5, t6, t7, - static_cast<const null_type&>(null_type()), - static_cast<const null_type&>(null_type())) {} - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2, - typename access_traits<T3>::parameter_type t3, - typename access_traits<T4>::parameter_type t4, - typename access_traits<T5>::parameter_type t5, - typename access_traits<T6>::parameter_type t6, - typename access_traits<T7>::parameter_type t7, - typename access_traits<T8>::parameter_type t8) - : inherited(t0, t1, t2, t3, t4, t5, t6, t7, t8, - static_cast<const null_type&>(null_type())) {} - - inline THRUST_HOST_DEVICE - tuple(typename access_traits<T0>::parameter_type t0, - typename access_traits<T1>::parameter_type t1, - typename access_traits<T2>::parameter_type t2, - typename access_traits<T3>::parameter_type t3, - typename access_traits<T4>::parameter_type t4, - typename access_traits<T5>::parameter_type t5, - typename access_traits<T6>::parameter_type t6, - typename access_traits<T7>::parameter_type t7, - typename access_traits<T8>::parameter_type t8, - typename access_traits<T9>::parameter_type t9) - : inherited(t0, t1, t2, t3, t4, t5, t6, t7, t8, t9) {} - - - template<class U1, class U2> - inline THRUST_HOST_DEVICE - tuple(const detail::cons<U1, U2>& p) : inherited(p) {} - - __thrust_exec_check_disable__ - template <class U1, class U2> - inline THRUST_HOST_DEVICE - tuple& operator=(const detail::cons<U1, U2>& k) - { - inherited::operator=(k); - return *this; - } - - /*! \endcond - */ - - /*! This assignment operator allows assigning the first two elements of this \p tuple from a \p pair. - * \param k A \p pair to assign from. - */ - __thrust_exec_check_disable__ - template <class U1, class U2> - THRUST_HOST_DEVICE inline - tuple& operator=(const thrust::pair<U1, U2>& k) { - //BOOST_STATIC_ASSERT(length<tuple>::value == 2);// check_length = 2 - this->head = k.first; - this->tail.head = k.second; - return *this; - } - - /*! \p swap swaps the elements of two <tt>tuple</tt>s. - * - * \param t The other <tt>tuple</tt> with which to swap. - */ - inline THRUST_HOST_DEVICE - void swap(tuple &t) - { - inherited::swap(t); - } -}; - -/*! \cond - */ - -template <> -class tuple<null_type, null_type, null_type, null_type, null_type, null_type, null_type, null_type, null_type, null_type> : - public null_type -{ -public: - typedef null_type inherited; -}; - -/*! \endcond - */ - - -/*! This version of \p make_tuple creates a new \c tuple object from a - * single object. - * - * \param t0 The object to copy from. - * \return A \p tuple object with a single member which is a copy of \p t0. - */ -template<class T0> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0>::type - make_tuple(const T0& t0); - -/*! This version of \p make_tuple creates a new \c tuple object from two - * objects. - * - * \param t0 The first object to copy from. - * \param t1 The second object to copy from. - * \return A \p tuple object with two members which are copies of \p t0 - * and \p t1. - * - * \note \p make_tuple has ten variants, the rest of which are omitted here - * for brevity. - */ -template<class T0, class T1> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1>::type - make_tuple(const T0& t0, const T1& t1); - -/*! This version of \p tie creates a new \c tuple whose single element is - * a reference which refers to this function's argument. - * - * \param t0 The object to reference. - * \return A \p tuple object with one member which is a reference to \p t0. - */ -template<typename T0> -THRUST_HOST_DEVICE inline -tuple<T0&> tie(T0& t0); - -/*! This version of \p tie creates a new \c tuple of references object which - * refers to this function's arguments. - * - * \param t0 The first object to reference. - * \param t1 The second object to reference. - * \return A \p tuple object with two members which are references to \p t0 - * and \p t1. - * - * \note \p tie has ten variants, the rest of which are omitted here for - * brevity. - */ -template<typename T0, typename T1> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&> tie(T0& t0, T1& t1); - -/*! \p swap swaps the contents of two <tt>tuple</tt>s. - * - * \param x The first \p tuple to swap. - * \param y The second \p tuple to swap. - */ -template< - typename T0, typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8, typename T9, - typename U0, typename U1, typename U2, typename U3, typename U4, typename U5, typename U6, typename U7, typename U8, typename U9 -> -inline THRUST_HOST_DEVICE -void swap(tuple<T0,T1,T2,T3,T4,T5,T6,T7,T8,T9> &x, - tuple<U0,U1,U2,U3,U4,U5,U6,U7,U8,U9> &y); - - - -/*! \cond - */ - -template<class T0, class T1, class T2> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2); - -template<class T0, class T1, class T2, class T3> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3); - -template<class T0, class T1, class T2, class T3, class T4> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4); - -template<class T0, class T1, class T2, class T3, class T4, class T5> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5); - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6); - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6, class T7> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6, T7>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6, const T7& t7); - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6, class T7, class T8> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6, T7, T8>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6, const T7& t7, const T8& t8); - -template<class T0, class T1, class T2, class T3, class T4, class T5, class T6, class T7, class T8, class T9> -THRUST_HOST_DEVICE inline - typename detail::make_tuple_mapper<T0, T1, T2, T3, T4, T5, T6, T7, T8, T9>::type - make_tuple(const T0& t0, const T1& t1, const T2& t2, const T3& t3, const T4& t4, const T5& t5, const T6& t6, const T7& t7, const T8& t8, const T9& t9); - -template<typename T0, typename T1, typename T2> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&> tie(T0 &t0, T1 &t1, T2 &t2); - -template<typename T0, typename T1, typename T2, typename T3> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3); - -template<typename T0, typename T1, typename T2, typename T3, typename T4> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4); - -template<typename T0, typename T1, typename T2, typename T3, typename T4, typename T5> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5); - -template<typename T0, typename T1, typename T2, typename T3, typename T4, typename T5, typename T6> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6); - -template<typename T0, typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6, T7 &t7); - -template<typename T0, typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&,T8&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6, T7 &t7, T8 &t8); - -template<typename T0, typename T1, typename T2, typename T3, typename T4, typename T5, typename T6, typename T7, typename T8, typename T9> -THRUST_HOST_DEVICE inline -tuple<T0&,T1&,T2&,T3&,T4&,T5&,T6&,T7&,T8&,T9&> tie(T0 &t0, T1 &t1, T2 &t2, T3 &t3, T4 &t4, T5 &t5, T6 &t6, T7 &t7, T8 &t8, T9 &t9); - - -THRUST_HOST_DEVICE inline -bool operator==(const null_type&, const null_type&); - -THRUST_HOST_DEVICE inline -bool operator>=(const null_type&, const null_type&); - -THRUST_HOST_DEVICE inline -bool operator<=(const null_type&, const null_type&); - -THRUST_HOST_DEVICE inline -bool operator!=(const null_type&, const null_type&); - -THRUST_HOST_DEVICE inline -bool operator<(const null_type&, const null_type&); +template <class... T> +using tuple = ::cuda::std::tuple<T...>; -THRUST_HOST_DEVICE inline -bool operator>(const null_type&, const null_type&); +using ::cuda::std::get; +using ::cuda::std::make_tuple; +using ::cuda::std::tie; /*! \endcond */ |