From e7ade77f4c9547fd53031feb72e9fca7f935d037 Mon Sep 17 00:00:00 2001 From: Michael Schellenberger Costa Date: Wed, 11 Sep 2024 13:35:35 +0200 Subject: [PATCH] Drop implementation of `thrust::pair` and `thrust::tuple` (#2395) * Drop implementation of `thrust::pair` and `thrust::tuple` We previously moved them back to proper class definitions, as using alias declarations broke CTAD. Thanks to @bernhardmgruber who realized that instead of making them an alias we can just pull them in and be done with it. Co-authored-by: Bernhard Manfred Gruber --- docs/thrust/api_docs/utility/pair.rst | 7 +- docs/thrust/api_docs/utility/tuple.rst | 6 +- .../cuda/std/detail/libcxx/include/tuple | 13 ++ .../detail/tuple_of_iterator_references.h | 13 +- thrust/thrust/pair.h | 98 ++----------- thrust/thrust/tuple.h | 138 +++--------------- 6 files changed, 60 insertions(+), 215 deletions(-) diff --git a/docs/thrust/api_docs/utility/pair.rst b/docs/thrust/api_docs/utility/pair.rst index 417dd9bcdc..e866f4849a 100644 --- a/docs/thrust/api_docs/utility/pair.rst +++ b/docs/thrust/api_docs/utility/pair.rst @@ -3,11 +3,12 @@ Pair ---- - - :cpp:struct:`thrust::pair ` +Thrust provides an implementations of `std::pair`, that is pulled in from `libcu++`. + +It is recommended to replace `thrust::pair` and `thrust::make_pair` with `cuda::std::pair` and `cuda::std::make_pair`. .. toctree:: :glob: :maxdepth: 1 - ${repo_docs_api_path}/*struct*pair* - ${repo_docs_api_path}/*function_group__pair* + ${repo_docs_api_path}/*typedef*pair* diff --git a/docs/thrust/api_docs/utility/tuple.rst b/docs/thrust/api_docs/utility/tuple.rst index d5d353f415..cab4babe32 100644 --- a/docs/thrust/api_docs/utility/tuple.rst +++ b/docs/thrust/api_docs/utility/tuple.rst @@ -3,10 +3,12 @@ Tuple ------ - - :cpp:struct:`thrust::tuple ` +Thrust provides an implementation of `std::tuple`, that is pulled in from `libcu++`. + +It is recommended to replace `thrust::tuple` with `cuda::std::tuple`. .. toctree:: :glob: :maxdepth: 1 - ${repo_docs_api_path}/*function_group__tuple* + ${repo_docs_api_path}/*typedef*tuple* diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/tuple b/libcudacxx/include/cuda/std/detail/libcxx/include/tuple index f8392629a4..5006082d59 100644 --- a/libcudacxx/include/cuda/std/detail/libcxx/include/tuple +++ b/libcudacxx/include/cuda/std/detail/libcxx/include/tuple @@ -194,6 +194,10 @@ _CCCL_PUSH_MACROS _LIBCUDACXX_BEGIN_NAMESPACE_STD +template +struct __is_tuple_of_iterator_references : _CUDA_VSTD::false_type +{}; + // __tuple_leaf struct __tuple_leaf_default_constructor_tag {}; @@ -854,6 +858,15 @@ public: typename __tuple_constraints<_Tp...>::template __tuple_like_constraints<_Tuple>, __invalid_tuple_constraints>; + // Horrible hack to make tuple_of_iterator_references work + template ::value, int> = 0, + __enable_if_t<(tuple_size<_TupleOfIteratorReferences>::value == sizeof...(_Tp)), int> = 0> + _LIBCUDACXX_HIDE_FROM_ABI _CCCL_CONSTEXPR_CXX14 tuple(_TupleOfIteratorReferences&& __t) + : tuple(_CUDA_VSTD::forward<_TupleOfIteratorReferences>(__t).template __to_tuple<_Tp...>( + __make_tuple_indices_t())) + {} + template , __enable_if_t::value, int> = 0, diff --git a/thrust/thrust/iterator/detail/tuple_of_iterator_references.h b/thrust/thrust/iterator/detail/tuple_of_iterator_references.h index befa2d3658..1c4e3d99ac 100644 --- a/thrust/thrust/iterator/detail/tuple_of_iterator_references.h +++ b/thrust/thrust/iterator/detail/tuple_of_iterator_references.h @@ -67,7 +67,9 @@ class tuple_of_iterator_references : public thrust::tuple using super_t = thrust::tuple; using super_t::super_t; - tuple_of_iterator_references() = default; + inline _CCCL_HOST_DEVICE tuple_of_iterator_references() + : super_t() + {} // allow implicit construction from tuple inline _CCCL_HOST_DEVICE tuple_of_iterator_references(const super_t& other) @@ -136,15 +138,14 @@ class tuple_of_iterator_references : public thrust::tuple } // namespace detail -template -struct __is_tuple_of_iterator_references> - : _CUDA_VSTD::true_type -{}; - THRUST_NAMESPACE_END _LIBCUDACXX_BEGIN_NAMESPACE_STD +template +struct __is_tuple_of_iterator_references> : true_type +{}; + // define tuple_size, tuple_element, etc. template struct tuple_size> diff --git a/thrust/thrust/pair.h b/thrust/thrust/pair.h index def1aeaf17..e3c7467799 100644 --- a/thrust/thrust/pair.h +++ b/thrust/thrust/pair.h @@ -30,9 +30,6 @@ # pragma system_header #endif // no system header -#include - -#include #include THRUST_NAMESPACE_BEGIN @@ -52,8 +49,12 @@ THRUST_NAMESPACE_BEGIN * \tparam N This parameter selects the member of interest. * \tparam T A \c pair type of interest. */ +#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen template using tuple_element = _CUDA_VSTD::tuple_element; +#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +using _CUDA_VSTD::tuple_element; +#endif // DOXYGEN_SHOULD_SKIP_THIS /*! This convenience metafunction is included for compatibility with * \p tuple. It returns \c 2, the number of elements of a \p pair, @@ -61,8 +62,12 @@ using tuple_element = _CUDA_VSTD::tuple_element; * * \tparam Pair A \c pair type of interest. */ +#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen template using tuple_size = _CUDA_VSTD::tuple_size; +#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +using _CUDA_VSTD::tuple_size; +#endif // DOXYGEN_SHOULD_SKIP_THIS /*! \p pair is a generic data structure encapsulating a heterogeneous * pair of values. @@ -75,55 +80,15 @@ using tuple_size = _CUDA_VSTD::tuple_size; * requirements on the type of \p T2. T2's type is * provided by pair::second_type. */ +#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen template -struct pair : public _CUDA_VSTD::pair -{ - using super_t = _CUDA_VSTD::pair; - using super_t::super_t; - -#if (defined(_CCCL_COMPILER_GCC) && __GNUC__ < 9) || (defined(_CCCL_COMPILER_CLANG) && __clang_major__ < 12) - // For whatever reason nvcc complains about that constructor being used before being defined in a constexpr variable - constexpr pair() = default; - - template ::template __constructible<_U1, _U2>, - _CUDA_VSTD::__enable_if_t<_Constraints::__implicit_constructible, int> = 0> - _CCCL_HOST_DEVICE constexpr pair(_U1&& __u1, _U2&& __u2) - : super_t(_CUDA_VSTD::forward<_U1>(__u1), _CUDA_VSTD::forward<_U2>(__u2)) - {} -#endif // _CCCL_COMPILER_GCC < 9 || _CCCL_COMPILER_CLANG < 12 -}; - -#if _CCCL_STD_VER >= 2017 -template -_CCCL_HOST_DEVICE pair(_T1, _T2) -> pair<_T1, _T2>; -#endif // _CCCL_STD_VER >= 2017 - -template -inline _CCCL_HOST_DEVICE -_CUDA_VSTD::__enable_if_t<_CUDA_VSTD::__is_swappable::value && _CUDA_VSTD::__is_swappable::value, void> -swap(pair& lhs, pair& rhs) noexcept( - (_CUDA_VSTD::__is_nothrow_swappable::value && _CUDA_VSTD::__is_nothrow_swappable::value)) -{ - lhs.swap(rhs); -} - -template -inline _CCCL_HOST_DEVICE -pair::type, typename _CUDA_VSTD::__unwrap_ref_decay::type> -make_pair(T1&& t1, T2&& t2) -{ - return pair::type, typename _CUDA_VSTD::__unwrap_ref_decay::type>( - _CUDA_VSTD::forward(t1), _CUDA_VSTD::forward(t2)); -} +using pair = _CUDA_VSTD::pair; +#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +using _CUDA_VSTD::pair; +#endif // DOXYGEN_SHOULD_SKIP_THIS using _CUDA_VSTD::get; - -template -struct proclaim_trivially_relocatable> - : ::cuda::std::conjunction, is_trivially_relocatable> -{}; +using _CUDA_VSTD::make_pair; /*! \endcond */ @@ -135,38 +100,3 @@ struct proclaim_trivially_relocatable> */ THRUST_NAMESPACE_END - -_LIBCUDACXX_BEGIN_NAMESPACE_STD - -template -struct tuple_size> : tuple_size> -{}; - -template -struct tuple_element> : tuple_element> -{}; - -template -struct __tuple_like_ext> : true_type -{}; - -_LIBCUDACXX_END_NAMESPACE_STD - -// This is a workaround for the fact that structured bindings require that the specializations of -// `tuple_size` and `tuple_element` reside in namespace std (https://eel.is/c++draft/dcl.struct.bind#4). -// See https://github.com/NVIDIA/libcudacxx/issues/316 for a short discussion -#if _CCCL_STD_VER >= 2017 - -# include - -namespace std -{ -template -struct tuple_size> : tuple_size> -{}; - -template -struct tuple_element> : tuple_element> -{}; -} // namespace std -#endif // _CCCL_STD_VER >= 2017 diff --git a/thrust/thrust/tuple.h b/thrust/thrust/tuple.h index 3e7cd87dae..1f8ed8943e 100644 --- a/thrust/thrust/tuple.h +++ b/thrust/thrust/tuple.h @@ -39,13 +39,7 @@ # pragma system_header #endif // no system header -#include - #include -#include -#include - -#include THRUST_NAMESPACE_BEGIN @@ -100,8 +94,12 @@ _CCCL_HOST_DEVICE inline bool operator>(const null_type&, const null_type&) * \see pair * \see tuple */ +#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen template using tuple_element = _CUDA_VSTD::tuple_element; +#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +using _CUDA_VSTD::tuple_element; +#endif // DOXYGEN_SHOULD_SKIP_THIS /*! This metafunction returns the number of elements * of a \p tuple type of interest. @@ -111,21 +109,19 @@ using tuple_element = _CUDA_VSTD::tuple_element; * \see pair * \see tuple */ +#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen template using tuple_size = _CUDA_VSTD::tuple_size; +#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +using _CUDA_VSTD::tuple_size; +#endif // DOXYGEN_SHOULD_SKIP_THIS -template -struct __is_tuple_of_iterator_references : _CUDA_VSTD::false_type -{}; - -/*! \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 - * tuple. Consequently, tuples are heterogeneous, fixed-size collections of - * values. An instantiation of \p tuple with two arguments is similar to an +/*! \brief \p tuple is a heterogeneous, fixed-size collection of values. + * An instantiation of \p tuple with two arguments is similar to an * instantiation of \p pair with the same two arguments. Individual elements * of a \p tuple may be accessed with the \p get function. * - * \tparam TN The type of the N \c tuple element. Thrust's \p tuple + * \tparam Ts The type of the N \c tuple element. Thrust's \p tuple * type currently supports up to ten elements. * * The following code snippet demonstrates how to create a new \p tuple object @@ -157,86 +153,16 @@ struct __is_tuple_of_iterator_references : _CUDA_VSTD::false_type * \see tuple_size * \see tie */ +#ifdef DOXYGEN_SHOULD_SKIP_THIS // Provide a fake alias for doxygen template -struct tuple : public _CUDA_VSTD::tuple -{ - using super_t = _CUDA_VSTD::tuple; - using super_t::super_t; - - tuple() = default; - - template ::value, int> = 0, - _CUDA_VSTD::__enable_if_t<(tuple_size<_TupleOfIteratorReferences>::value == sizeof...(Ts)), int> = 0> - _CCCL_HOST_DEVICE tuple(_TupleOfIteratorReferences&& tup) - : tuple(_CUDA_VSTD::forward<_TupleOfIteratorReferences>(tup).template __to_tuple( - _CUDA_VSTD::__make_tuple_indices_t())) - {} - - _CCCL_EXEC_CHECK_DISABLE - template ::value, int> = 0> - _CCCL_HOST_DEVICE tuple& operator=(TupleLike&& other) - { - super_t::operator=(_CUDA_VSTD::forward(other)); - return *this; - } - -#if defined(_CCCL_COMPILER_MSVC_2017) - // MSVC2017 needs some help to convert tuples - template , tuple>::value, int> = 0, - _CUDA_VSTD::__enable_if_t<_CUDA_VSTD::__tuple_convertible<_CUDA_VSTD::tuple, super_t>::value, int> = 0> - _CCCL_HOST_DEVICE constexpr operator tuple() - { - return __to_tuple(typename _CUDA_VSTD::__make_tuple_indices::type{}); - } - - template - _CCCL_HOST_DEVICE constexpr tuple __to_tuple(_CUDA_VSTD::__tuple_indices) const - { - return tuple{_CUDA_VSTD::get(*this)...}; - } -#endif // _CCCL_COMPILER_MSVC_2017 -}; - -#if _CCCL_STD_VER >= 2017 -template -_CCCL_HOST_DEVICE tuple(Ts...) -> tuple; - -template -struct pair; - -template -_CCCL_HOST_DEVICE tuple(pair) -> tuple; -#endif // _CCCL_STD_VER >= 2017 - -template -inline - _CCCL_HOST_DEVICE _CUDA_VSTD::__enable_if_t<_CUDA_VSTD::__all<_CUDA_VSTD::__is_swappable::value...>::value, void> - swap(tuple& __x, - tuple& __y) noexcept((_CUDA_VSTD::__all<_CUDA_VSTD::__is_nothrow_swappable::value...>::value)) -{ - __x.swap(__y); -} - -template -inline _CCCL_HOST_DEVICE tuple::type...> make_tuple(Ts&&... __t) -{ - return tuple::type...>(_CUDA_VSTD::forward(__t)...); -} - -template -inline _CCCL_HOST_DEVICE tuple tie(Ts&... ts) noexcept -{ - return tuple(ts...); -} +using tuple = _CUDA_VSTD::tuple; +#else // ^^^ DOXYGEN_SHOULD_SKIP_THIS ^^^ / vvv !DOXYGEN_SHOULD_SKIP_THIS vvv +using _CUDA_VSTD::tuple; +#endif // DOXYGEN_SHOULD_SKIP_THIS using _CUDA_VSTD::get; - -template -struct proclaim_trivially_relocatable> : ::cuda::std::conjunction...> -{}; +using _CUDA_VSTD::make_tuple; +using _CUDA_VSTD::tie; /*! \endcond */ @@ -251,18 +177,6 @@ THRUST_NAMESPACE_END _LIBCUDACXX_BEGIN_NAMESPACE_STD -template -struct tuple_size> : tuple_size> -{}; - -template -struct tuple_element> : tuple_element> -{}; - -template -struct __tuple_like_ext> : true_type -{}; - template <> struct tuple_size= 2017 -namespace std -{ -template -struct tuple_size> : tuple_size> -{}; - -template -struct tuple_element> : tuple_element> -{}; -} // namespace std -#endif // _CCCL_STD_VER >= 2017