Ensure that cuda::std::pair is potentially trivially copyable (#1249)

trivially copyable is a requirement for memcpy. We want to ensure that our pair implementation satisfies that whenever possible. This is especially important for thrust::pair as that is used in rmm extensively. Fixes #1246 Co-authored-by: Georgy Evtushenko <[email protected]>
NVIDIA · Jan 18, 2024 · a4a3289 · a4a3289
1 parent 5243a84
commit a4a3289
Show file tree

Hide file tree

Showing 13 changed files with 656 additions and 835 deletions.
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__config b/libcudacxx/include/cuda/std/detail/libcxx/include/__config
@@ -184,16 +184,6 @@ extern "C++" {
 // conflict with the dllexport-emitted copy, so we disable it.
 #    define _LIBCUDACXX_DEPRECATED_ABI_LEGACY_LIBRARY_DEFINITIONS_FOR_INLINE_FUNCTIONS
 #  endif
-// Feature macros for disabling pre ABI v1 features. All of these options
-// are deprecated.
-#  if defined(__FreeBSD__)
-#    define _LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR
-#  endif
-#endif
-
-#ifdef _LIBCUDACXX_TRIVIAL_PAIR_COPY_CTOR
-#error "_LIBCUDACXX_TRIVIAL_PAIR_COPY_CTOR" is no longer supported. \
-       use _LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR instead
 #endif
 
 #ifndef __has_attribute

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__utility/pair.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__utility/pair.h
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/tuple b/libcudacxx/include/cuda/std/detail/libcxx/include/tuple
@@ -1327,16 +1327,17 @@ template <class... _Tp, class _Alloc>
 struct _LIBCUDACXX_TEMPLATE_VIS uses_allocator<tuple<_Tp...>, _Alloc>
     : true_type {};
 
-template <class _T1, class _T2>
+template <class _T1, class _T2, bool _IsRef>
 template <class... _Args1, class... _Args2, size_t... _I1, size_t... _I2>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX17
-pair<_T1, _T2>::pair(piecewise_construct_t, tuple<_Args1...> &__first_args,
-                     tuple<_Args2...> &__second_args, __tuple_indices<_I1...>,
-                     __tuple_indices<_I2...>)
-    : first(_CUDA_VSTD::forward<_Args1>(_CUDA_VSTD::get<_I1>(__first_args))...),
-      second(
-          _CUDA_VSTD::forward<_Args2>(_CUDA_VSTD::get<_I2>(__second_args))...) {
-}
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX17 __pair_base<_T1, _T2, _IsRef>::__pair_base(
+  piecewise_construct_t,
+  tuple<_Args1...>& __first_args,
+  tuple<_Args2...>& __second_args,
+  __tuple_indices<_I1...>,
+  __tuple_indices<_I2...>)
+    : first(_CUDA_VSTD::forward<_Args1>(_CUDA_VSTD::get<_I1>(__first_args))...)
+    , second(_CUDA_VSTD::forward<_Args2>(_CUDA_VSTD::get<_I2>(__second_args))...)
+{}
 
 #if _CCCL_STD_VER > 2014
 #define _LIBCUDACXX_NOEXCEPT_RETURN(...)                                       \

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/type_traits b/libcudacxx/include/cuda/std/detail/libcxx/include/type_traits
@@ -423,6 +423,7 @@ namespace std
 #endif // __cuda_std__
 
 #include "__assert" // all public C++ headers provide the assertion handler
+#include "__fwd/pair.h"
 #include "__functional/identity.h"
 #include "__functional/invoke.h"
 #include "__memory/addressof.h"
@@ -571,7 +572,6 @@ namespace std
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template <class _T1, class _T2> struct _LIBCUDACXX_TEMPLATE_VIS pair;
 template <class _Tp> class _LIBCUDACXX_TEMPLATE_VIS reference_wrapper;
 template <class _Tp> struct _LIBCUDACXX_TEMPLATE_VIS hash;
 

diff --git a/.../libcxx/test/libcxx/utilities/utility/pairs/pairs.pair/non_trivial_copy_move_ABI.pass.cpp b/.../libcxx/test/libcxx/utilities/utility/pairs/pairs.pair/non_trivial_copy_move_ABI.pass.cpp
diff --git a/...acxx/libcxx/test/libcxx/utilities/utility/pairs/pairs.pair/trivial_copy_move_ABI.pass.cpp b/...acxx/libcxx/test/libcxx/utilities/utility/pairs/pairs.pair/trivial_copy_move_ABI.pass.cpp
@@ -26,10 +26,6 @@
 
 #include "test_macros.h"
 
-#if defined(_LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR)
-#error Non-trivial ctor ABI macro defined
-#endif
-
 template <class T>
 struct HasTrivialABI : std::integral_constant<bool,
     std::is_trivially_destructible<T>::value

diff --git a/libcudacxx/libcxx/test/std/utilities/utility/pairs/pairs.pair/trivial_copy_move.pass.cpp b/libcudacxx/libcxx/test/std/utilities/utility/pairs/pairs.pair/trivial_copy_move.pass.cpp
@@ -28,25 +28,19 @@ int main(int, char**)
     typedef std::pair<int, short> P;
     {
         static_assert(std::is_copy_constructible<P>::value, "");
-#if !defined(_LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR)
         static_assert(std::is_trivially_copy_constructible<P>::value, "");
-#endif
     }
 #if TEST_STD_VER >= 2011
     {
         static_assert(std::is_move_constructible<P>::value, "");
-#if !defined(_LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR)
         static_assert(std::is_trivially_move_constructible<P>::value, "");
-#endif
     }
     {
         using P1 = std::pair<Dummy, int>;
         static_assert(!std::is_copy_constructible<P1>::value, "");
         static_assert(!std::is_trivially_copy_constructible<P1>::value, "");
         static_assert(std::is_move_constructible<P1>::value, "");
-#if !defined(_LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR)
         static_assert(std::is_trivially_move_constructible<P1>::value, "");
-#endif
     }
 #endif
 

diff --git a/libcudacxx/test/libcudacxx/std/utilities/optional/optional.object/special_members.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/optional/optional.object/special_members.pass.cpp
@@ -9,9 +9,6 @@
 
 // UNSUPPORTED: c++03, c++11
 
-// UNSUPPORTED: nvrtc
-// see nvbug4263883
-
 // <cuda/std/optional>
 
 // Make sure we properly generate special member functions for optional<T>

diff --git a/libcudacxx/test/libcudacxx/std/utilities/optional/optional.object/triviality.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/optional/optional.object/triviality.pass.cpp
@@ -9,9 +9,6 @@
 
 // UNSUPPORTED: c++03, c++11
 
-// UNSUPPORTED: nvrtc
-// see nvbug4263883
-
 // <cuda/std/optional>
 
 // The following special member functions should propagate the triviality of

diff --git a/libcudacxx/test/libcudacxx/std/utilities/utility/pairs/pairs.pair/trivial_copy_move.pass.cpp b/libcudacxx/test/libcudacxx/std/utilities/utility/pairs/pairs.pair/trivial_copy_move.pass.cpp
@@ -30,24 +30,31 @@ int main(int, char**)
     typedef cuda::std::pair<int, short> P;
     {
         static_assert(cuda::std::is_copy_constructible<P>::value, "");
-#if !defined(_LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR)
         static_assert(cuda::std::is_trivially_copy_constructible<P>::value, "");
-#endif
     }
     {
         static_assert(cuda::std::is_move_constructible<P>::value, "");
-#if !defined(_LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR)
         static_assert(cuda::std::is_trivially_move_constructible<P>::value, "");
-#endif
     }
     {
         using P1 = cuda::std::pair<Dummy, int>;
         static_assert(!cuda::std::is_copy_constructible<P1>::value, "");
         static_assert(!cuda::std::is_trivially_copy_constructible<P1>::value, "");
         static_assert(cuda::std::is_move_constructible<P1>::value, "");
-#if !defined(_LIBCUDACXX_DEPRECATED_ABI_DISABLE_PAIR_TRIVIAL_COPY_CTOR)
         static_assert(cuda::std::is_trivially_move_constructible<P1>::value, "");
-#endif
+    }
+
+    // extensions to ensure pair is trivially_copyable
+    {
+        static_assert(cuda::std::is_copy_assignable<P>::value, "");
+        static_assert(cuda::std::is_trivially_copy_assignable<P>::value, "");
+    }
+    {
+        static_assert(cuda::std::is_move_assignable<P>::value, "");
+        static_assert(cuda::std::is_trivially_move_assignable<P>::value, "");
+    }
+    {
+        static_assert(cuda::std::is_trivially_copyable<P>::value, "");
     }
 
   return 0;

diff --git a/libcudacxx/test/support/archetypes.ipp b/libcudacxx/test/support/archetypes.ipp
@@ -24,6 +24,13 @@
 #define DEFINE_ASSIGN_CONSTEXPR
 #endif
 #endif
+#ifndef DEFINE_DEFAULT_CONSTEXPR
+#if defined(TEST_COMPILER_NVRTC)
+#define DEFINE_DEFAULT_CONSTEXPR
+#else
+#define DEFINE_DEFAULT_CONSTEXPR DEFINE_CONSTEXPR
+#endif
+#endif
 #ifndef DEFINE_CTOR
 #define DEFINE_CTOR = default
 #undef  DEFINE_INIT_LIST // defaulted constructors do not require explicit initializers for the base class
@@ -59,7 +66,7 @@ struct AllCtors : DEFINE_BASE(AllCtors) {
   using Base::Base;
 #endif
   using Base::operator=;
-  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR AllCtors() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
+  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_DEFAULT_CONSTEXPR AllCtors() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
   DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR AllCtors(AllCtors const&) DEFINE_NOEXCEPT DEFINE_INIT_LIST DEFINE_CTOR;
   DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR AllCtors(AllCtors &&) DEFINE_NOEXCEPT DEFINE_INIT_LIST DEFINE_CTOR;
   DEFINE_ASSIGN_ANNOTATIONS DEFINE_ASSIGN_CONSTEXPR AllCtors& operator=(AllCtors const&) DEFINE_NOEXCEPT DEFINE_ASSIGN;
@@ -101,7 +108,7 @@ struct DefaultOnly : DEFINE_BASE(DefaultOnly) {
 #else
   using Base::Base;
 #endif
-  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR DefaultOnly() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
+  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_DEFAULT_CONSTEXPR DefaultOnly() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
   DefaultOnly(DefaultOnly const&) DEFINE_NOEXCEPT = delete;
   DefaultOnly& operator=(DefaultOnly const&) DEFINE_NOEXCEPT = delete;
   DEFINE_DTOR(DefaultOnly)
@@ -118,7 +125,7 @@ struct Copyable : DEFINE_BASE(Copyable) {
 #else
   using Base::Base;
 #endif
-  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR Copyable() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
+  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_DEFAULT_CONSTEXPR Copyable() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
   DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR Copyable(Copyable const &) DEFINE_NOEXCEPT DEFINE_INIT_LIST DEFINE_CTOR;
   DEFINE_ASSIGN_ANNOTATIONS Copyable &operator=(Copyable const &) DEFINE_NOEXCEPT DEFINE_ASSIGN;
   DEFINE_DTOR(Copyable)
@@ -135,7 +142,7 @@ struct CopyOnly : DEFINE_BASE(CopyOnly) {
 #else
   using Base::Base;
 #endif
-  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR CopyOnly() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
+  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_DEFAULT_CONSTEXPR CopyOnly() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
   DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR CopyOnly(CopyOnly const &) DEFINE_NOEXCEPT DEFINE_INIT_LIST DEFINE_CTOR;
   DEFINE_EXPLICIT DEFINE_CONSTEXPR CopyOnly(CopyOnly &&) DEFINE_NOEXCEPT = delete;
   DEFINE_ASSIGN_ANNOTATIONS CopyOnly &operator=(CopyOnly const &) DEFINE_NOEXCEPT DEFINE_ASSIGN;
@@ -154,7 +161,7 @@ struct NonCopyable : DEFINE_BASE(NonCopyable) {
 #else
   using Base::Base;
 #endif
-  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR NonCopyable() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
+  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_DEFAULT_CONSTEXPR NonCopyable() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
   DEFINE_EXPLICIT DEFINE_CONSTEXPR NonCopyable(NonCopyable const &) DEFINE_NOEXCEPT = delete;
   NonCopyable &operator=(NonCopyable const &) DEFINE_NOEXCEPT = delete;
   DEFINE_DTOR(NonCopyable)
@@ -171,7 +178,7 @@ struct MoveOnly : DEFINE_BASE(MoveOnly) {
 #else
   using Base::Base;
 #endif
-  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR MoveOnly() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
+  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_DEFAULT_CONSTEXPR MoveOnly() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
   DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR MoveOnly(MoveOnly &&) DEFINE_NOEXCEPT DEFINE_INIT_LIST DEFINE_CTOR;
   DEFINE_ASSIGN_ANNOTATIONS MoveOnly &operator=(MoveOnly &&) DEFINE_NOEXCEPT DEFINE_ASSIGN;
   DEFINE_DTOR(MoveOnly)
@@ -233,7 +240,7 @@ struct ConvertingType : DEFINE_BASE(ConvertingType) {
 #else
   using Base::Base;
 #endif
-  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR ConvertingType() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
+  DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_DEFAULT_CONSTEXPR ConvertingType() DEFINE_NOEXCEPT DEFINE_DEFAULT_CTOR;
   DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR ConvertingType(ConvertingType const&) DEFINE_NOEXCEPT DEFINE_INIT_LIST DEFINE_CTOR;
   DEFINE_CTOR_ANNOTATIONS DEFINE_EXPLICIT DEFINE_CONSTEXPR ConvertingType(ConvertingType &&) DEFINE_NOEXCEPT DEFINE_INIT_LIST DEFINE_CTOR;
   DEFINE_ASSIGN_ANNOTATIONS ConvertingType& operator=(ConvertingType const&) DEFINE_NOEXCEPT DEFINE_ASSIGN;