From 44e6418f2227a3aeb55a6dc917d8dbe30e08ce20 Mon Sep 17 00:00:00 2001
From: Michael Schellenberger Costa <miscco@nvidia.com>
Date: Mon, 9 Oct 2023 15:03:35 +0200
Subject: [PATCH] Modernize tuple

---
 .../tuple.assign/convert_copy.pass.cpp        |    2 -
 .../tuple.assign/convert_move.pass.cpp        |    2 -
 .../tuple.tuple/tuple.assign/copy.pass.cpp    |    2 -
 .../tuple.tuple/tuple.assign/move.pass.cpp    |    2 -
 ...PR22806_constrain_tuple_like_ctor.pass.cpp |    8 -
 .../tuple.cnstr/alloc_move.pass.cpp           |    6 -
 .../tuple.cnstr/convert_copy.pass.cpp         |    2 -
 .../tuple.tuple/tuple.cnstr/move.pass.cpp     |    4 -
 .../tuple.cnstr/nothrow_cnstr.pass.cpp        |    4 -
 .../tuple.creation/forward_as_tuple.pass.cpp  |    2 -
 .../tuple.creation/tuple_cat.pass.cpp         |    8 -
 .../tuple.elem/get_const_rv.pass.cpp          |    2 -
 .../tuple.tuple/tuple.elem/get_rv.pass.cpp    |    2 -
 .../tuple.elem/tuple.by.type.pass.cpp         |    2 -
 .../tuple.special/non_member_swap.pass.cpp    |    2 -
 .../tuple.swap/member_swap.pass.cpp           |    2 -
 .../detail/libcxx/include/__functional/bind.h |   13 +-
 .../std/detail/libcxx/include/__fwd/get.h     |   68 +-
 .../libcxx/include/__tuple_dir/apply_cv.h     |   60 +-
 .../include/__tuple_dir/make_tuple_types.h    |   49 +-
 .../include/__tuple_dir/sfinae_helpers.h      |  187 +-
 .../include/__tuple_dir/structured_bindings.h |  279 ++-
 .../include/__tuple_dir/tuple_element.h       |   83 +-
 .../include/__tuple_dir/tuple_indices.h       |   11 +-
 .../libcxx/include/__tuple_dir/tuple_like.h   |   46 +-
 .../libcxx/include/__tuple_dir/tuple_size.h   |   54 +-
 .../libcxx/include/__tuple_dir/tuple_types.h  |    8 +-
 .../detail/libcxx/include/__utility/pair.h    |   20 +-
 .../cuda/std/detail/libcxx/include/tuple      | 1946 ++++++++---------
 libcudacxx/libcxx/docs/UsingLibcxx.rst        |   27 -
 ...ed_arity_initialization_extension.pass.cpp |  119 -
 31 files changed, 1327 insertions(+), 1695 deletions(-)
 delete mode 100644 libcudacxx/libcxx/test/libcxx/utilities/tuple/tuple.tuple/tuple.cnstr/enable_reduced_arity_initialization_extension.pass.cpp

diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_copy.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_copy.pass.cpp
index 2a2c00858b..403f744a71 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_copy.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_copy.pass.cpp
@@ -75,7 +75,6 @@ int main(int, char**)
         assert(cuda::std::get<1>(t1) == int('a'));
         assert(cuda::std::get<2>(t1).id_ == 2);
     }
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         // Test that tuple evaluates correctly applies an lvalue reference
         // before evaluating is_assignable (ie 'is_assignable<int&, int&>')
@@ -88,6 +87,5 @@ int main(int, char**)
         assert(cuda::std::get<0>(t) == 43);
         assert(&cuda::std::get<0>(t) == &x);
     }
-#endif
   return 0;
 }
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_move.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_move.pass.cpp
index 77f97ad884..741e79e953 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_move.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/convert_move.pass.cpp
@@ -99,7 +99,6 @@ int main(int, char**)
         assert(cuda::std::get<1>(t1) == int('a'));
         assert(cuda::std::get<2>(t1)->id_ == 3);
     }*/
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         // Test that tuple evaluates correctly applies an lvalue reference
         // before evaluating is_assignable (ie 'is_assignable<int&, int&&>')
@@ -112,6 +111,5 @@ int main(int, char**)
         assert(cuda::std::get<0>(t) == 43);
         assert(&cuda::std::get<0>(t) == &x);
     }
-#endif
   return 0;
 }
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/copy.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/copy.pass.cpp
index 65ea14ca88..d9b5464999 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/copy.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/copy.pass.cpp
@@ -72,7 +72,6 @@ int main(int, char**)
         assert(cuda::std::get<2>(t) == "some text");
     }
     */
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         // test reference assignment.
         using T = cuda::std::tuple<int&, int&&>;
@@ -88,7 +87,6 @@ int main(int, char**)
         assert(cuda::std::get<1>(t) == y2);
         assert(&cuda::std::get<1>(t) == &y);
     }
-#endif
     // cuda::std::unique_ptr not supported
     /*
     {
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/move.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/move.pass.cpp
index 0d0af9fc12..5f138ba09a 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/move.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.assign/move.pass.cpp
@@ -57,7 +57,6 @@ int main(int, char**)
         t = cuda::std::move(t0);
         unused(t);
     }
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         typedef cuda::std::tuple<MoveOnly> T;
         T t0(MoveOnly(0));
@@ -97,7 +96,6 @@ int main(int, char**)
         assert(cuda::std::get<1>(t) == y2);
         assert(&cuda::std::get<1>(t) == &y);
     }
-#endif
     // cuda::std::unique_ptr not supported
     /*
     {
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/PR22806_constrain_tuple_like_ctor.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/PR22806_constrain_tuple_like_ctor.pass.cpp
index 7c4eac6003..3c1ebf84cd 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/PR22806_constrain_tuple_like_ctor.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/PR22806_constrain_tuple_like_ctor.pass.cpp
@@ -97,11 +97,9 @@ int main(int, char**)
     // cuda::std::allocator not supported
     // cuda::std::allocator<int> A;
     { // rvalue reference
-#if !defined(TEST_COMPILER_MSVC_2017)
         T t1(42);
         cuda::std::tuple< T&& > t2(cuda::std::move(t1));
         assert(&cuda::std::get<0>(t2) == &t1);
-#endif
     }
     { // const lvalue reference
         T t1(42);
@@ -119,12 +117,10 @@ int main(int, char**)
         assert(&cuda::std::get<0>(t2) == &t1);
     }
     { // const rvalue reference
-#if !defined(TEST_COMPILER_MSVC_2017)
         T t1(42);
 
         cuda::std::tuple< T const && > t2(cuda::std::move(t1));
         assert(&cuda::std::get<0>(t2) == &t1);
-#endif
     }
     // cuda::std::allocator not supported
     /*
@@ -159,20 +155,16 @@ int main(int, char**)
     // the 'tuple(UTypes...)' ctor should be chosen and 'UDT' constructed from
     // 'tuple<T>'.
     {
-#if !defined(TEST_COMPILER_MSVC_2017)
         using VT = ConstructibleFromTupleAndInt;
         cuda::std::tuple<int> t1(42);
         cuda::std::tuple<VT> t2(t1);
         assert(cuda::std::get<0>(t2).state == VT::FromTuple);
-#endif
     }
     {
-#if !defined(TEST_COMPILER_MSVC_2017)
         using VT = ConvertibleFromTupleAndInt;
         cuda::std::tuple<int> t1(42);
         cuda::std::tuple<VT> t2 = {t1};
         assert(cuda::std::get<0>(t2).state == VT::FromTuple);
-#endif
     }
     // Test constructing a 1-tuple of the form tuple<UDT> from another 1-tuple
     // 'tuple<T>' where UDT cannot be constructed from 'tuple<T>' but can
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/alloc_move.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/alloc_move.pass.cpp
index 771a5cc355..87ca0bb701 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/alloc_move.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/alloc_move.pass.cpp
@@ -33,12 +33,10 @@ int main(int, char**)
         T t(cuda::std::allocator_arg, A1<int>(), cuda::std::move(t0));
     }
     {
-#if !defined(TEST_COMPILER_MSVC_2017)
         typedef cuda::std::tuple<MoveOnly> T;
         T t0(MoveOnly(0));
         T t(cuda::std::allocator_arg, A1<int>(), cuda::std::move(t0));
         assert(cuda::std::get<0>(t) == 0);
-#endif
     }
     {
         typedef cuda::std::tuple<alloc_first> T;
@@ -59,7 +57,6 @@ int main(int, char**)
 // testing extensions
 #ifdef _LIBCUDACXX_VERSION
     {
-#if !defined(TEST_COMPILER_MSVC_2017)
         typedef cuda::std::tuple<MoveOnly, alloc_first> T;
         T t0(0 ,1);
         alloc_first::allocator_constructed() = false;
@@ -67,10 +64,8 @@ int main(int, char**)
         assert(alloc_first::allocator_constructed());
         assert(cuda::std::get<0>(t) == 0);
         assert(cuda::std::get<1>(t) == 1);
-#endif
     }
     {
-#if !defined(TEST_COMPILER_MSVC_2017)
         typedef cuda::std::tuple<MoveOnly, alloc_first, alloc_last> T;
         T t0(1, 2, 3);
         alloc_first::allocator_constructed() = false;
@@ -81,7 +76,6 @@ int main(int, char**)
         assert(cuda::std::get<0>(t) == 1);
         assert(cuda::std::get<1>(t) == 2);
         assert(cuda::std::get<2>(t) == 3);
-#endif
     }
 #endif
 
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/convert_copy.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/convert_copy.pass.cpp
index 9cc3cf3406..0ca9d28c66 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/convert_copy.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/convert_copy.pass.cpp
@@ -150,11 +150,9 @@ int main(int, char**)
         static_assert(cuda::std::is_convertible<ExplicitTwo&&, ExplicitTwo>::value, "");
         static_assert(cuda::std::is_convertible<cuda::std::tuple<ExplicitTwo&&>&&, const cuda::std::tuple<ExplicitTwo>&>::value, "");
 
-#if !defined(TEST_COMPILER_MSVC_2017)
         ExplicitTwo e;
         cuda::std::tuple<ExplicitTwo> t = cuda::std::tuple<ExplicitTwo&&>(cuda::std::move(e));
         ((void)t);
-#endif
     }
   return 0;
 }
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/move.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/move.pass.cpp
index baf5a7d6c2..97419419e8 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/move.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/move.pass.cpp
@@ -64,9 +64,7 @@ __host__ __device__ void test_sfinae() {
     }
     // args constructors
     {
-#if !defined(TEST_COMPILER_MSVC_2017)
         static_assert(cuda::std::is_constructible<Tup, Elem&&>::value, "");
-#endif
         static_assert(!cuda::std::is_constructible<Tup, Elem const&>::value, "");
         static_assert(!cuda::std::is_constructible<Tup, Elem&>::value, "");
     }
@@ -96,7 +94,6 @@ int main(int, char**)
         T t = cuda::std::move(t0);
         unused(t); // Prevent unused warning
     }
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         typedef cuda::std::tuple<MoveOnly> T;
         T t0(MoveOnly(0));
@@ -127,7 +124,6 @@ int main(int, char**)
         d_t d2(static_cast<d_t &&>(d));
         unused(d2);
     }
-#endif
     {
         test_sfinae<move_only_ebo>();
         test_sfinae<move_only_large>();
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/nothrow_cnstr.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/nothrow_cnstr.pass.cpp
index 5e3ca56c02..5a140d8b19 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/nothrow_cnstr.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.cnstr/nothrow_cnstr.pass.cpp
@@ -14,10 +14,6 @@
 
 // UNSUPPORTED: c++98, c++03
 
-
-// XFAIL: gcc-8 && c++17 && !nvrtc
-// XFAIL: gcc-7 && c++17 && !nvrtc
-
 #include <cuda/std/tuple>
 #include <cuda/std/utility>
 #include <cuda/std/cassert>
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/forward_as_tuple.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/forward_as_tuple.pass.cpp
index f51a130f36..30dac05be8 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/forward_as_tuple.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/forward_as_tuple.pass.cpp
@@ -67,11 +67,9 @@ int main(int, char**)
     {
         test0(cuda::std::forward_as_tuple());
     }
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         test1a(cuda::std::forward_as_tuple(1));
     }
-#endif
     {
         int i = 2;
         test1b(cuda::std::forward_as_tuple(i));
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/tuple_cat.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/tuple_cat.pass.cpp
index c4970eb21b..047f8d22ee 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/tuple_cat.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.creation/tuple_cat.pass.cpp
@@ -78,7 +78,6 @@ int main(int, char**)
         unused(t); // Prevent unused warning
     }
     */
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         constexpr cuda::std::tuple<int> t1(1);
         constexpr cuda::std::tuple<int> t = cuda::std::tuple_cat(t1);
@@ -91,15 +90,12 @@ int main(int, char**)
         static_assert(cuda::std::get<1>(t) == 1, "");
     }
 #endif
-#endif
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         cuda::std::tuple<int, MoveOnly> t =
                                 cuda::std::tuple_cat(cuda::std::tuple<int, MoveOnly>(1, 2));
         assert(cuda::std::get<0>(t) == 1);
         assert(cuda::std::get<1>(t) == 2);
     }
-#endif
     // cuda::std::array not supported
     /*
     {
@@ -109,13 +105,11 @@ int main(int, char**)
         assert(cuda::std::get<2>(t) == 0);
     }
     */
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         cuda::std::tuple<int, MoveOnly> t = cuda::std::tuple_cat(cuda::std::pair<int, MoveOnly>(2, 1));
         assert(cuda::std::get<0>(t) == 2);
         assert(cuda::std::get<1>(t) == 1);
     }
-#endif
     {
         cuda::std::tuple<> t1;
         cuda::std::tuple<> t2;
@@ -164,7 +158,6 @@ int main(int, char**)
         assert(cuda::std::get<1>(t3) == 3.5);
         assert(cuda::std::get<2>(t3) == nullptr);
     }
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         cuda::std::tuple<int*, MoveOnly> t1(nullptr, 1);
         cuda::std::tuple<int, double> t2(2, 3.5);
@@ -275,6 +268,5 @@ int main(int, char**)
             int, const int, int&, const int&>);
         unused(r);
     }
-#endif
   return 0;
 }
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_const_rv.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_const_rv.pass.cpp
index 7ef111245d..124a9164dc 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_const_rv.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_const_rv.pass.cpp
@@ -65,7 +65,6 @@ int main(int, char**)
     static_assert(noexcept(cuda::std::get<1>(cuda::std::move(p))), "");
     }
 
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
     int x = 42;
     int const y = 43;
@@ -75,7 +74,6 @@ int main(int, char**)
     static_assert(cuda::std::is_same<int const&&, decltype(cuda::std::get<1>(cuda::std::move(p)))>::value, "");
     static_assert(noexcept(cuda::std::get<1>(cuda::std::move(p))), "");
     }
-#endif
 
 #if TEST_STD_VER > 11
     {
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_rv.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_rv.pass.cpp
index f48d59da4d..100876d305 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_rv.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/get_rv.pass.cpp
@@ -37,12 +37,10 @@ int main(int, char**)
         assert(*p == 3);
     }
     */
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         cuda::std::tuple<MoveOnly> t(3);
         MoveOnly _m = cuda::std::get<0>(cuda::std::move(t));
         assert(_m.get() == 3);
     }
-#endif
   return 0;
 }
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/tuple.by.type.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/tuple.by.type.pass.cpp
index dca715e004..dde1b18e32 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/tuple.by.type.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.elem/tuple.by.type.pass.cpp
@@ -89,7 +89,6 @@ int main(int, char**)
     static_assert(noexcept(cuda::std::get<int const&>(cuda::std::move(t))), "");
     }
 
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
     int x = 42;
     int y = 43;
@@ -99,7 +98,6 @@ int main(int, char**)
     static_assert(cuda::std::is_same<int const&&, decltype(cuda::std::get<int const&&>(cuda::std::move(t)))>::value, "");
     static_assert(noexcept(cuda::std::get<int const&&>(cuda::std::move(t))), "");
     }
-#endif
     {
     constexpr const cuda::std::tuple<int, const int, double, double> t { 1, 2, 3.4, 5.6 };
     static_assert(cuda::std::get<int>(cuda::std::move(t)) == 1, "");
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.special/non_member_swap.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.special/non_member_swap.pass.cpp
index 2d0533a912..7a1119764b 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.special/non_member_swap.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.special/non_member_swap.pass.cpp
@@ -30,7 +30,6 @@ int main(int, char**)
         T t1;
         swap(t0, t1);
     }
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         typedef cuda::std::tuple<MoveOnly> T;
         T t0(MoveOnly(0));
@@ -61,6 +60,5 @@ int main(int, char**)
         assert(cuda::std::get<1>(t1) == 1);
         assert(cuda::std::get<2>(t1) == 2);
     }
-#endif
   return 0;
 }
diff --git a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.swap/member_swap.pass.cpp b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.swap/member_swap.pass.cpp
index e5b3f2f394..1db35af6e2 100644
--- a/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.swap/member_swap.pass.cpp
+++ b/libcudacxx/.upstream-tests/test/std/utilities/tuple/tuple.tuple/tuple.swap/member_swap.pass.cpp
@@ -29,7 +29,6 @@ int main(int, char**)
         T t1;
         t0.swap(t1);
     }
-#if !defined(TEST_COMPILER_MSVC_2017)
     {
         typedef cuda::std::tuple<MoveOnly> T;
         T t0(MoveOnly(0));
@@ -60,6 +59,5 @@ int main(int, char**)
         assert(cuda::std::get<1>(t1) == 1);
         assert(cuda::std::get<2>(t1) == 2);
     }
-#endif
   return 0;
 }
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__functional/bind.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__functional/bind.h
index f0b62ff4bc..c90edb8143 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__functional/bind.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__functional/bind.h
@@ -132,7 +132,7 @@ __enable_if_t
 >
 __mu(_Ti& __ti, tuple<_Uj...>& __uj)
 {
-    typedef typename __make_tuple_indices<sizeof...(_Uj)>::type __indices;
+    typedef __make_tuple_indices_t<sizeof...(_Uj)> __indices;
     return _CUDA_VSTD::__mu_expand(__ti, __uj, __indices());
 }
 
@@ -142,7 +142,7 @@ struct __mu_return2 {};
 template <class _Ti, class _Uj>
 struct __mu_return2<true, _Ti, _Uj>
 {
-    typedef typename tuple_element<is_placeholder<_Ti>::value - 1, _Uj>::type type;
+  typedef __tuple_element_t<is_placeholder<_Ti>::value - 1, _Uj> type;
 };
 
 template <class _Ti, class _Uj>
@@ -154,8 +154,8 @@ __enable_if_t
 >
 __mu(_Ti&, _Uj& __uj)
 {
-    const size_t _Indx = is_placeholder<_Ti>::value - 1;
-    return _CUDA_VSTD::forward<typename tuple_element<_Indx, _Uj>::type>(_CUDA_VSTD::get<_Indx>(__uj));
+  const size_t _Indx = is_placeholder<_Ti>::value - 1;
+  return _CUDA_VSTD::forward<__tuple_element_t<_Indx, _Uj>>(_CUDA_VSTD::get<_Indx>(__uj));
 }
 
 template <class _Ti, class _Uj>
@@ -197,8 +197,7 @@ struct __mu_return_impl<_Ti, false, true, false, tuple<_Uj...> >
 template <class _Ti, class _TupleUj>
 struct __mu_return_impl<_Ti, false, false, true, _TupleUj>
 {
-    typedef typename tuple_element<is_placeholder<_Ti>::value - 1,
-                                   _TupleUj>::type&& type;
+  typedef __tuple_element_t<is_placeholder<_Ti>::value - 1, _TupleUj>&& type;
 };
 
 template <class _Ti, class _TupleUj>
@@ -298,7 +297,7 @@ class __bind : public __weak_result_type<__decay_t<_Fp>>
     _Fd __f_;
     _Td __bound_args_;
 
-    typedef typename __make_tuple_indices<sizeof...(_BoundArgs)>::type __indices;
+    typedef __make_tuple_indices_t<sizeof...(_BoundArgs)> __indices;
 public:
     template <class _Gp, class ..._BA,
               class = __enable_if_t
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__fwd/get.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__fwd/get.h
index fd18cf672f..69b3a53fe1 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__fwd/get.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__fwd/get.h
@@ -11,7 +11,7 @@
 #define _LIBCUDACXX___FWD_GET_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__fwd/array.h"
@@ -21,70 +21,54 @@
 #include "../cstddef"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template <size_t _Ip, class ..._Tp>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, tuple<_Tp...> >::type&
-get(tuple<_Tp...>&) noexcept;
+template <size_t _Ip, class... _Tp>
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, tuple<_Tp...>>& get(
+    tuple<_Tp...>&) noexcept;
 
-template <size_t _Ip, class ..._Tp>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, tuple<_Tp...> >::type&
-get(const tuple<_Tp...>&) noexcept;
+template <size_t _Ip, class... _Tp>
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, tuple<_Tp...>>& get(
+    const tuple<_Tp...>&) noexcept;
 
-template <size_t _Ip, class ..._Tp>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, tuple<_Tp...> >::type&&
-get(tuple<_Tp...>&&) noexcept;
+template <size_t _Ip, class... _Tp>
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, tuple<_Tp...>>&& get(
+    tuple<_Tp...>&&) noexcept;
 
-template <size_t _Ip, class ..._Tp>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, tuple<_Tp...> >::type&&
-get(const tuple<_Tp...>&&) noexcept;
+template <size_t _Ip, class... _Tp>
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, tuple<_Tp...>>&& get(
+    const tuple<_Tp...>&&) noexcept;
 
 template <size_t _Ip, class _T1, class _T2>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, pair<_T1, _T2> >::type&
-get(pair<_T1, _T2>&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, pair<_T1, _T2>>& get(
+    pair<_T1, _T2>&) noexcept;
 
 template <size_t _Ip, class _T1, class _T2>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, pair<_T1, _T2> >::type&
-get(const pair<_T1, _T2>&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, pair<_T1, _T2>>& get(
+    const pair<_T1, _T2>&) noexcept;
 
 template <size_t _Ip, class _T1, class _T2>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, pair<_T1, _T2> >::type&&
-get(pair<_T1, _T2>&&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, pair<_T1, _T2>>&& get(
+    pair<_T1, _T2>&&) noexcept;
 
 template <size_t _Ip, class _T1, class _T2>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, pair<_T1, _T2> >::type&&
-get(const pair<_T1, _T2>&&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, pair<_T1, _T2>>&& get(
+    const pair<_T1, _T2>&&) noexcept;
 
 template <size_t _Ip, class _Tp, size_t _Size>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-_Tp&
-get(array<_Tp, _Size>&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _Tp& get(array<_Tp, _Size>&) noexcept;
 
 template <size_t _Ip, class _Tp, size_t _Size>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const _Tp&
-get(const array<_Tp, _Size>&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const _Tp& get(const array<_Tp, _Size>&) noexcept;
 
 template <size_t _Ip, class _Tp, size_t _Size>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-_Tp&&
-get(array<_Tp, _Size>&&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _Tp&& get(array<_Tp, _Size>&&) noexcept;
 
 template <size_t _Ip, class _Tp, size_t _Size>
-_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const _Tp&&
-get(const array<_Tp, _Size>&&) noexcept;
+_LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const _Tp&& get(const array<_Tp, _Size>&&) noexcept;
 
 _LIBCUDACXX_END_NAMESPACE_STD
 
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/apply_cv.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/apply_cv.h
index 06de3942f9..ff670542dd 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/apply_cv.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/apply_cv.h
@@ -11,7 +11,7 @@
 #define _LIBCUDACXX___TUPLE_APPLY_CV_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__type_traits/is_const.h"
@@ -20,7 +20,7 @@
 #include "../__type_traits/remove_reference.h"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
@@ -28,42 +28,58 @@ _LIBCUDACXX_BEGIN_NAMESPACE_STD
 template <bool _ApplyLV, bool _ApplyConst, bool _ApplyVolatile>
 struct __apply_cv_mf;
 template <>
-struct __apply_cv_mf<false, false, false> {
-  template <class _Tp> using __apply = _Tp;
+struct __apply_cv_mf<false, false, false>
+{
+  template <class _Tp>
+  using __apply = _Tp;
 };
 template <>
-struct __apply_cv_mf<false, true, false> {
-  template <class _Tp> using __apply _LIBCUDACXX_NODEBUG_TYPE = const _Tp;
+struct __apply_cv_mf<false, true, false>
+{
+  template <class _Tp>
+  using __apply _LIBCUDACXX_NODEBUG_TYPE = const _Tp;
 };
 template <>
-struct __apply_cv_mf<false, false, true> {
-  template <class _Tp> using __apply _LIBCUDACXX_NODEBUG_TYPE = volatile _Tp;
+struct __apply_cv_mf<false, false, true>
+{
+  template <class _Tp>
+  using __apply _LIBCUDACXX_NODEBUG_TYPE = volatile _Tp;
 };
 template <>
-struct __apply_cv_mf<false, true, true> {
-  template <class _Tp> using __apply _LIBCUDACXX_NODEBUG_TYPE = const volatile _Tp;
+struct __apply_cv_mf<false, true, true>
+{
+  template <class _Tp>
+  using __apply _LIBCUDACXX_NODEBUG_TYPE = const volatile _Tp;
 };
 template <>
-struct __apply_cv_mf<true, false, false> {
-  template <class _Tp> using __apply _LIBCUDACXX_NODEBUG_TYPE = _Tp&;
+struct __apply_cv_mf<true, false, false>
+{
+  template <class _Tp>
+  using __apply _LIBCUDACXX_NODEBUG_TYPE = _Tp&;
 };
 template <>
-struct __apply_cv_mf<true, true, false> {
-  template <class _Tp> using __apply _LIBCUDACXX_NODEBUG_TYPE = const _Tp&;
+struct __apply_cv_mf<true, true, false>
+{
+  template <class _Tp>
+  using __apply _LIBCUDACXX_NODEBUG_TYPE = const _Tp&;
 };
 template <>
-struct __apply_cv_mf<true, false, true> {
-  template <class _Tp> using __apply _LIBCUDACXX_NODEBUG_TYPE = volatile _Tp&;
+struct __apply_cv_mf<true, false, true>
+{
+  template <class _Tp>
+  using __apply _LIBCUDACXX_NODEBUG_TYPE = volatile _Tp&;
 };
 template <>
-struct __apply_cv_mf<true, true, true> {
-  template <class _Tp> using __apply _LIBCUDACXX_NODEBUG_TYPE = const volatile _Tp&;
+struct __apply_cv_mf<true, true, true>
+{
+  template <class _Tp>
+  using __apply _LIBCUDACXX_NODEBUG_TYPE = const volatile _Tp&;
 };
 template <class _Tp, class _RawTp = __libcpp_remove_reference_t<_Tp> >
-using __apply_cv_t _LIBCUDACXX_NODEBUG_TYPE = __apply_cv_mf<
-    is_lvalue_reference<_Tp>::value,
-    is_const<_RawTp>::value,
-    is_volatile<_RawTp>::value>;
+using __apply_cv_t _LIBCUDACXX_NODEBUG_TYPE =
+    __apply_cv_mf<_LIBCUDACXX_TRAIT(is_lvalue_reference, _Tp),
+                  _LIBCUDACXX_TRAIT(is_const, _RawTp),
+                  _LIBCUDACXX_TRAIT(is_volatile, _RawTp)>;
 
 _LIBCUDACXX_END_NAMESPACE_STD
 
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/make_tuple_types.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/make_tuple_types.h
index ba9dbb9e80..6c3a4d1482 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/make_tuple_types.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/make_tuple_types.h
@@ -11,7 +11,7 @@
 #define _LIBCUDACXX___TUPLE_MAKE_TUPLE_TYPES_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__fwd/array.h"
@@ -26,7 +26,7 @@
 #include "../cstddef"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
@@ -39,46 +39,49 @@ _LIBCUDACXX_BEGIN_NAMESPACE_STD
 template <class _TupleTypes, class _TupleIndices>
 struct __make_tuple_types_flat;
 
-template <template <class...> class _Tuple, class ..._Types, size_t ..._Idx>
-struct __make_tuple_types_flat<_Tuple<_Types...>, __tuple_indices<_Idx...>> {
+template <template <class...> class _Tuple, class... _Types, size_t... _Idx>
+struct __make_tuple_types_flat<_Tuple<_Types...>, __tuple_indices<_Idx...>>
+{
   // Specialization for pair, tuple, and __tuple_types
   template <class _Tp, class _ApplyFn = __apply_cv_t<_Tp>>
-  using __apply_quals _LIBCUDACXX_NODEBUG_TYPE = __tuple_types<
-      typename _ApplyFn::template __apply<__type_pack_element<_Idx, _Types...>>...
-    >;
+  using __apply_quals _LIBCUDACXX_NODEBUG_TYPE =
+      __tuple_types< typename _ApplyFn::template __apply<__type_pack_element<_Idx, _Types...>>... >;
 };
 
-template <class _Vt, size_t _Np, size_t ..._Idx>
-struct __make_tuple_types_flat<array<_Vt, _Np>, __tuple_indices<_Idx...>> {
+template <class _Vt, size_t _Np, size_t... _Idx>
+struct __make_tuple_types_flat<array<_Vt, _Np>, __tuple_indices<_Idx...>>
+{
   template <size_t>
   using __value_type = _Vt;
   template <class _Tp, class _ApplyFn = __apply_cv_t<_Tp>>
-  using __apply_quals = __tuple_types<
-      typename _ApplyFn::template __apply<__value_type<_Idx>>...
-    >;
+  using __apply_quals = __tuple_types< typename _ApplyFn::template __apply<__value_type<_Idx>>... >;
 };
 
-template <class _Tp, size_t _Ep = tuple_size<__libcpp_remove_reference_t<_Tp> >::value,
-          size_t _Sp = 0,
-          bool _SameSize = (_Ep == tuple_size<__libcpp_remove_reference_t<_Tp> >::value)>
+template <class _Tp, size_t _Ep = tuple_size<__libcpp_remove_reference_t<_Tp>>::value, size_t _Sp = 0,
+    bool _SameSize = (_Ep == tuple_size<__libcpp_remove_reference_t<_Tp>>::value)>
 struct __make_tuple_types
 {
-    static_assert(_Sp <= _Ep, "__make_tuple_types input error");
-    using _RawTp = __remove_cv_t<__libcpp_remove_reference_t<_Tp>>;
-    using _Maker = __make_tuple_types_flat<_RawTp, typename __make_tuple_indices<_Ep, _Sp>::type>;
-    using type = typename _Maker::template __apply_quals<_Tp>;
+  static_assert(_Sp <= _Ep, "__make_tuple_types input error");
+  using _RawTp = __remove_cv_t<__libcpp_remove_reference_t<_Tp>>;
+  using _Maker = __make_tuple_types_flat<_RawTp, __make_tuple_indices_t<_Ep, _Sp>>;
+  using type   = typename _Maker::template __apply_quals<_Tp>;
 };
 
-template <class ..._Types, size_t _Ep>
-struct __make_tuple_types<tuple<_Types...>, _Ep, 0, true> {
+template <class... _Types, size_t _Ep>
+struct __make_tuple_types<tuple<_Types...>, _Ep, 0, true>
+{
   typedef _LIBCUDACXX_NODEBUG_TYPE __tuple_types<_Types...> type;
 };
 
-template <class ..._Types, size_t _Ep>
-struct __make_tuple_types<__tuple_types<_Types...>, _Ep, 0, true> {
+template <class... _Types, size_t _Ep>
+struct __make_tuple_types<__tuple_types<_Types...>, _Ep, 0, true>
+{
   typedef _LIBCUDACXX_NODEBUG_TYPE __tuple_types<_Types...> type;
 };
 
+template <class _Tp, size_t _Ep = tuple_size<__libcpp_remove_reference_t<_Tp>>::value, size_t _Sp = 0>
+using __make_tuple_types_t = typename __make_tuple_types<_Tp, _Ep, _Sp>::type;
+
 _LIBCUDACXX_END_NAMESPACE_STD
 
 #endif // _LIBCUDACXX___TUPLE_MAKE_TUPLE_TYPES_H
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/sfinae_helpers.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/sfinae_helpers.h
index 5662850d5d..77993f05b5 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/sfinae_helpers.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/sfinae_helpers.h
@@ -11,7 +11,7 @@
 #define _LIBCUDACXX___TUPLE_SFINAE_HELPERS_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__fwd/tuple.h"
@@ -31,33 +31,35 @@
 #include "../cstddef"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template <bool ..._Preds>
+template <bool... _Preds>
 struct __all_dummy;
 
-template <bool ..._Pred>
-using __all = _IsSame<__all_dummy<_Pred...>, __all_dummy<((void)_Pred, true)...>>;
+template <bool... _Pred>
+using __all = _IsSame<__all_dummy<_Pred...>, __all_dummy<((void) _Pred, true)...>>;
 
-struct __tuple_sfinae_base {
+struct __tuple_sfinae_base
+{
   template <class, class>
-  struct __test_size : false_type {};
+  struct __test_size : false_type
+  {};
 
-  template <class ..._Tp, class ..._Up>
-  struct __test_size<__tuple_types<_Tp...>, __tuple_types<_Up...>>
-    : _BoolConstant<sizeof...(_Tp) == sizeof...(_Up)> {};
+  template <class... _Tp, class... _Up>
+  struct __test_size<__tuple_types<_Tp...>, __tuple_types<_Up...>> : _BoolConstant<sizeof...(_Tp) == sizeof...(_Up)>
+  {};
 
-  template <template <class, class...> class,
-            class _Tp, class _Up, bool = __test_size<_Tp, _Up>::value>
-  struct __test : false_type {};
+  template <template <class, class...> class, class _Tp, class _Up, bool = __test_size<_Tp, _Up>::value>
+  struct __test : false_type
+  {};
 
-  template <template <class, class...> class _Trait,
-            class ..._LArgs, class ..._RArgs>
+  template <template <class, class...> class _Trait, class... _LArgs, class... _RArgs>
   struct __test<_Trait, __tuple_types<_LArgs...>, __tuple_types<_RArgs...>, true>
-      : __all<_Trait<_LArgs, _RArgs>::value...> {};
+      : __all<_Trait<_LArgs, _RArgs>::value...>
+  {};
 
   template <class _FromArgs, class _ToArgs>
   using __constructible = __test<is_constructible, _ToArgs, _FromArgs>;
@@ -70,136 +72,133 @@ struct __tuple_sfinae_base {
 // __tuple_convertible
 
 template <class _Tp, class _Up, bool = __tuple_like<__libcpp_remove_reference_t<_Tp>>::value,
-                                bool = __tuple_like<_Up>::value>
-struct __tuple_convertible
-    : public false_type {};
+    bool = __tuple_like<_Up>::value>
+struct __tuple_convertible : public false_type
+{};
 
 template <class _Tp, class _Up>
 struct __tuple_convertible<_Tp, _Up, true, true>
-    : public __tuple_sfinae_base::__convertible<
-      typename __make_tuple_types<_Tp>::type
-    , typename __make_tuple_types<_Up>::type
-    >
+    : public __tuple_sfinae_base::__convertible<__make_tuple_types_t<_Tp>, __make_tuple_types_t<_Up>>
 {};
 
 // __tuple_constructible
 
 template <class _Tp, class _Up, bool = __tuple_like<__libcpp_remove_reference_t<_Tp>>::value,
-                                bool = __tuple_like<_Up>::value>
-struct __tuple_constructible
-    : public false_type {};
+    bool = __tuple_like<_Up>::value>
+struct __tuple_constructible : public false_type
+{};
 
 template <class _Tp, class _Up>
 struct __tuple_constructible<_Tp, _Up, true, true>
-    : public __tuple_sfinae_base::__constructible<
-      typename __make_tuple_types<_Tp>::type
-    , typename __make_tuple_types<_Up>::type
-    >
+    : public __tuple_sfinae_base::__constructible<__make_tuple_types_t<_Tp>, __make_tuple_types_t<_Up>>
 {};
 
 // __tuple_assignable
 
 template <class _Tp, class _Up, bool = __tuple_like<__libcpp_remove_reference_t<_Tp>>::value,
-                                bool = __tuple_like<_Up>::value>
-struct __tuple_assignable
-    : public false_type {};
+    bool = __tuple_like<_Up>::value>
+struct __tuple_assignable : public false_type
+{};
 
 template <class _Tp, class _Up>
 struct __tuple_assignable<_Tp, _Up, true, true>
-    : public __tuple_sfinae_base::__assignable<
-      typename __make_tuple_types<_Tp>::type
-    , typename __make_tuple_types<_Up&>::type
-    >
+    : public __tuple_sfinae_base::__assignable<__make_tuple_types_t<_Tp>, __make_tuple_types_t<_Up&>>
 {};
 
-
-template <size_t _Ip, class ..._Tp>
+template <size_t _Ip, class... _Tp>
 struct _LIBCUDACXX_TEMPLATE_VIS tuple_element<_Ip, tuple<_Tp...> >
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename tuple_element<_Ip, __tuple_types<_Tp...> >::type type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE __tuple_element_t<_Ip, __tuple_types<_Tp...>> type;
 };
 
 template <bool _IsTuple, class _SizeTrait, size_t _Expected>
-struct __tuple_like_with_size_imp : false_type {};
+struct __tuple_like_with_size_imp : false_type
+{};
 
 template <class _SizeTrait, size_t _Expected>
-struct __tuple_like_with_size_imp<true, _SizeTrait, _Expected>
-    : integral_constant<bool, _SizeTrait::value == _Expected> {};
-
-template <class _Tuple, size_t _ExpectedSize,
-          class _RawTuple = __remove_cvref_t<_Tuple>>
-using __tuple_like_with_size _LIBCUDACXX_NODEBUG_TYPE = __tuple_like_with_size_imp<
-                                   __tuple_like<_RawTuple>::value,
-                                   tuple_size<_RawTuple>, _ExpectedSize
-                              >;
-
-struct _LIBCUDACXX_TYPE_VIS __check_tuple_constructor_fail {
-    template <int&...>
-    using __enable_explicit_default = false_type;
-    template <int&...>
-    using __enable_implicit_default = false_type;
-    template <class ...>
-    using __enable_explicit = false_type;
-    template <class ...>
-    using __enable_implicit = false_type;
-    template <class ...>
-    using __enable_assign = false_type;
+struct __tuple_like_with_size_imp<true, _SizeTrait, _Expected> : integral_constant<bool, _SizeTrait::value == _Expected>
+{};
+
+template <class _Tuple, size_t _ExpectedSize, class _RawTuple = __remove_cvref_t<_Tuple>>
+using __tuple_like_with_size _LIBCUDACXX_NODEBUG_TYPE =
+    __tuple_like_with_size_imp< __tuple_like<_RawTuple>::value, tuple_size<_RawTuple>, _ExpectedSize >;
+
+struct _LIBCUDACXX_TYPE_VIS __check_tuple_constructor_fail
+{
+  template <int&...>
+  using __enable_explicit_default = false_type;
+  template <int&...>
+  using __enable_implicit_default = false_type;
+  template <class...>
+  using __enable_explicit = false_type;
+  template <class...>
+  using __enable_implicit = false_type;
+  template <class...>
+  using __enable_assign = false_type;
 };
 
 #if _LIBCUDACXX_STD_VER > 11
 
 template <bool _CanCopy, bool _CanMove>
-struct __sfinae_ctor_base {};
+struct __sfinae_ctor_base
+{};
 template <>
-struct __sfinae_ctor_base<false, false> {
-  __sfinae_ctor_base() = default;
-  __sfinae_ctor_base(__sfinae_ctor_base const&) = delete;
-  __sfinae_ctor_base(__sfinae_ctor_base &&) = delete;
+struct __sfinae_ctor_base<false, false>
+{
+  __sfinae_ctor_base()                                     = default;
+  __sfinae_ctor_base(__sfinae_ctor_base const&)            = delete;
+  __sfinae_ctor_base(__sfinae_ctor_base&&)                 = delete;
   __sfinae_ctor_base& operator=(__sfinae_ctor_base const&) = default;
-  __sfinae_ctor_base& operator=(__sfinae_ctor_base&&) = default;
+  __sfinae_ctor_base& operator=(__sfinae_ctor_base&&)      = default;
 };
 template <>
-struct __sfinae_ctor_base<true, false> {
-  __sfinae_ctor_base() = default;
-  __sfinae_ctor_base(__sfinae_ctor_base const&) = default;
-  __sfinae_ctor_base(__sfinae_ctor_base &&) = delete;
+struct __sfinae_ctor_base<true, false>
+{
+  __sfinae_ctor_base()                                     = default;
+  __sfinae_ctor_base(__sfinae_ctor_base const&)            = default;
+  __sfinae_ctor_base(__sfinae_ctor_base&&)                 = delete;
   __sfinae_ctor_base& operator=(__sfinae_ctor_base const&) = default;
-  __sfinae_ctor_base& operator=(__sfinae_ctor_base&&) = default;
+  __sfinae_ctor_base& operator=(__sfinae_ctor_base&&)      = default;
 };
 template <>
-struct __sfinae_ctor_base<false, true> {
-  __sfinae_ctor_base() = default;
-  __sfinae_ctor_base(__sfinae_ctor_base const&) = delete;
-  __sfinae_ctor_base(__sfinae_ctor_base &&) = default;
+struct __sfinae_ctor_base<false, true>
+{
+  __sfinae_ctor_base()                                     = default;
+  __sfinae_ctor_base(__sfinae_ctor_base const&)            = delete;
+  __sfinae_ctor_base(__sfinae_ctor_base&&)                 = default;
   __sfinae_ctor_base& operator=(__sfinae_ctor_base const&) = default;
-  __sfinae_ctor_base& operator=(__sfinae_ctor_base&&) = default;
+  __sfinae_ctor_base& operator=(__sfinae_ctor_base&&)      = default;
 };
 
 template <bool _CanCopy, bool _CanMove>
-struct __sfinae_assign_base {};
+struct __sfinae_assign_base
+{};
 template <>
-struct __sfinae_assign_base<false, false> {
-  __sfinae_assign_base() = default;
-  __sfinae_assign_base(__sfinae_assign_base const&) = default;
-  __sfinae_assign_base(__sfinae_assign_base &&) = default;
+struct __sfinae_assign_base<false, false>
+{
+  __sfinae_assign_base()                                       = default;
+  __sfinae_assign_base(__sfinae_assign_base const&)            = default;
+  __sfinae_assign_base(__sfinae_assign_base&&)                 = default;
   __sfinae_assign_base& operator=(__sfinae_assign_base const&) = delete;
-  __sfinae_assign_base& operator=(__sfinae_assign_base&&) = delete;
+  __sfinae_assign_base& operator=(__sfinae_assign_base&&)      = delete;
 };
 template <>
-struct __sfinae_assign_base<true, false> {
-  __sfinae_assign_base() = default;
-  __sfinae_assign_base(__sfinae_assign_base const&) = default;
-  __sfinae_assign_base(__sfinae_assign_base &&) = default;
+struct __sfinae_assign_base<true, false>
+{
+  __sfinae_assign_base()                                       = default;
+  __sfinae_assign_base(__sfinae_assign_base const&)            = default;
+  __sfinae_assign_base(__sfinae_assign_base&&)                 = default;
   __sfinae_assign_base& operator=(__sfinae_assign_base const&) = default;
-  __sfinae_assign_base& operator=(__sfinae_assign_base&&) = delete;
+  __sfinae_assign_base& operator=(__sfinae_assign_base&&)      = delete;
 };
 template <>
-struct __sfinae_assign_base<false, true> {
-  __sfinae_assign_base() = default;
-  __sfinae_assign_base(__sfinae_assign_base const&) = default;
-  __sfinae_assign_base(__sfinae_assign_base &&) = default;
+struct __sfinae_assign_base<false, true>
+{
+  __sfinae_assign_base()                                       = default;
+  __sfinae_assign_base(__sfinae_assign_base const&)            = default;
+  __sfinae_assign_base(__sfinae_assign_base&&)                 = default;
   __sfinae_assign_base& operator=(__sfinae_assign_base const&) = delete;
-  __sfinae_assign_base& operator=(__sfinae_assign_base&&) = default;
+  __sfinae_assign_base& operator=(__sfinae_assign_base&&)      = default;
 };
 #endif // _LIBCUDACXX_STD_VER > 11
 
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/structured_bindings.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/structured_bindings.h
index 50768cdd2e..dd6b92daeb 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/structured_bindings.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/structured_bindings.h
@@ -12,163 +12,148 @@
 
 #ifdef __cuda_std__
 
-#if defined(_LIBCUDACXX_COMPILER_CLANG)
-#pragma clang diagnostic push
-#pragma clang diagnostic ignored "-Wmismatched-tags"
-#endif // _LIBCUDACXX_COMPILER_CLANG
+#  if defined(_LIBCUDACXX_COMPILER_CLANG)
+#    pragma clang diagnostic push
+#    pragma clang diagnostic ignored "-Wmismatched-tags"
+#  endif // _LIBCUDACXX_COMPILER_CLANG
 
-#if !defined(__CUDACC_RTC__)
+#  if !defined(__CUDACC_RTC__)
 // Fetch utility to get primary template for ::std::tuple_size necessary for the specialization of
 // ::std::tuple_size<cuda::std::tuple> to enable structured bindings.
 // See https://github.com/NVIDIA/libcudacxx/issues/316
-#include <utility>
-#endif
+#    include <utility>
+#  endif
 
-#include "../__fwd/array.h"
-#include "../__fwd/pair.h"
-#include "../__fwd/tuple.h"
-#include "../__tuple_dir/tuple_element.h"
-#include "../__tuple_dir/tuple_size.h"
-#include "../__type_traits/integral_constant.h"
+#  include "../__fwd/array.h"
+#  include "../__fwd/pair.h"
+#  include "../__fwd/tuple.h"
+#  include "../__tuple_dir/tuple_element.h"
+#  include "../__tuple_dir/tuple_size.h"
+#  include "../__type_traits/integral_constant.h"
 
 // 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 _LIBCUDACXX_STD_VER > 14
+#  if _LIBCUDACXX_STD_VER > 14
 namespace std {
-#if defined(__CUDACC_RTC__)
-    template <class... _Tp>
-    struct tuple_size;
-
-    template<size_t _Ip, class... _Tp>
-    struct tuple_element;
-#endif
-
-    template <class _Tp, size_t _Size>
-    struct tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template <class _Tp, size_t _Size>
-    struct tuple_size<const _CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template <class _Tp, size_t _Size>
-    struct tuple_size<volatile _CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template <class _Tp, size_t _Size>
-    struct tuple_size<const volatile _CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template<size_t _Ip, class _Tp, size_t _Size>
-    struct tuple_element<_Ip, _CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_element<_Ip, _CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template<size_t _Ip, class _Tp, size_t _Size>
-    struct tuple_element<_Ip, const _CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_element<_Ip, const _CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template<size_t _Ip, class _Tp, size_t _Size>
-    struct tuple_element<_Ip, volatile _CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_element<_Ip, volatile _CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template<size_t _Ip, class _Tp, size_t _Size>
-    struct tuple_element<_Ip, const volatile _CUDA_VSTD::array<_Tp, _Size>>
-      : _CUDA_VSTD::tuple_element<_Ip, const volatile _CUDA_VSTD::array<_Tp, _Size>>
-    {};
-
-    template <class _Tp, class _Up>
-    struct tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template <class _Tp, class _Up>
-    struct tuple_size<const _CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template <class _Tp, class _Up>
-    struct tuple_size<volatile _CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template <class _Tp, class _Up>
-    struct tuple_size<const volatile _CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template<size_t _Ip, class _Tp, class _Up>
-    struct tuple_element<_Ip, _CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_element<_Ip, _CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template<size_t _Ip, class _Tp, class _Up>
-    struct tuple_element<_Ip, const _CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_element<_Ip, const _CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template<size_t _Ip, class _Tp, class _Up>
-    struct tuple_element<_Ip, volatile _CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_element<_Ip, volatile _CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template<size_t _Ip, class _Tp, class _Up>
-    struct tuple_element<_Ip, const volatile _CUDA_VSTD::pair<_Tp, _Up>>
-      : _CUDA_VSTD::tuple_element<_Ip, const volatile _CUDA_VSTD::pair<_Tp, _Up>>
-    {};
-
-    template <class... _Tp>
-    struct tuple_size<_CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
-    {};
-
-    template <class... _Tp>
-    struct tuple_size<const _CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
-    {};
-
-    template <class... _Tp>
-    struct tuple_size<volatile _CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
-    {};
-
-    template <class... _Tp>
-    struct tuple_size<const volatile _CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
-    {};
-
-    template<size_t _Ip, class... _Tp>
-    struct tuple_element<_Ip, _CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_element<_Ip, _CUDA_VSTD::tuple<_Tp...>>
-    {};
-
-    template<size_t _Ip, class... _Tp>
-    struct tuple_element<_Ip, const _CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_element<_Ip, const _CUDA_VSTD::tuple<_Tp...>>
-    {};
-
-    template<size_t _Ip, class... _Tp>
-    struct tuple_element<_Ip, volatile _CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_element<_Ip, volatile _CUDA_VSTD::tuple<_Tp...>>
-    {};
-
-    template<size_t _Ip, class... _Tp>
-    struct tuple_element<_Ip, const volatile _CUDA_VSTD::tuple<_Tp...>>
-      : _CUDA_VSTD::tuple_element<_Ip, const volatile _CUDA_VSTD::tuple<_Tp...>>
-    {};
-}
-#endif // _LIBCUDACXX_STD_VER > 14
-
-#if defined(_LIBCUDACXX_COMPILER_CLANG)
-#pragma clang diagnostic pop
-# endif // _LIBCUDACXX_COMPILER_CLANG
+#    if defined(__CUDACC_RTC__)
+template <class... _Tp>
+struct tuple_size;
+
+template <size_t _Ip, class... _Tp>
+struct tuple_element;
+#    endif
+
+template <class _Tp, size_t _Size>
+struct tuple_size<_CUDA_VSTD::array<_Tp, _Size>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <class _Tp, size_t _Size>
+struct tuple_size<const _CUDA_VSTD::array<_Tp, _Size>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <class _Tp, size_t _Size>
+struct tuple_size<volatile _CUDA_VSTD::array<_Tp, _Size>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <class _Tp, size_t _Size>
+struct tuple_size<const volatile _CUDA_VSTD::array<_Tp, _Size>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <size_t _Ip, class _Tp, size_t _Size>
+struct tuple_element<_Ip, _CUDA_VSTD::array<_Tp, _Size>> : _CUDA_VSTD::tuple_element<_Ip, _CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <size_t _Ip, class _Tp, size_t _Size>
+struct tuple_element<_Ip, const _CUDA_VSTD::array<_Tp, _Size>>
+    : _CUDA_VSTD::tuple_element<_Ip, const _CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <size_t _Ip, class _Tp, size_t _Size>
+struct tuple_element<_Ip, volatile _CUDA_VSTD::array<_Tp, _Size>>
+    : _CUDA_VSTD::tuple_element<_Ip, volatile _CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <size_t _Ip, class _Tp, size_t _Size>
+struct tuple_element<_Ip, const volatile _CUDA_VSTD::array<_Tp, _Size>>
+    : _CUDA_VSTD::tuple_element<_Ip, const volatile _CUDA_VSTD::array<_Tp, _Size>>
+{};
+
+template <class _Tp, class _Up>
+struct tuple_size<_CUDA_VSTD::pair<_Tp, _Up>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <class _Tp, class _Up>
+struct tuple_size<const _CUDA_VSTD::pair<_Tp, _Up>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <class _Tp, class _Up>
+struct tuple_size<volatile _CUDA_VSTD::pair<_Tp, _Up>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <class _Tp, class _Up>
+struct tuple_size<const volatile _CUDA_VSTD::pair<_Tp, _Up>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <size_t _Ip, class _Tp, class _Up>
+struct tuple_element<_Ip, _CUDA_VSTD::pair<_Tp, _Up>> : _CUDA_VSTD::tuple_element<_Ip, _CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <size_t _Ip, class _Tp, class _Up>
+struct tuple_element<_Ip, const _CUDA_VSTD::pair<_Tp, _Up>>
+    : _CUDA_VSTD::tuple_element<_Ip, const _CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <size_t _Ip, class _Tp, class _Up>
+struct tuple_element<_Ip, volatile _CUDA_VSTD::pair<_Tp, _Up>>
+    : _CUDA_VSTD::tuple_element<_Ip, volatile _CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <size_t _Ip, class _Tp, class _Up>
+struct tuple_element<_Ip, const volatile _CUDA_VSTD::pair<_Tp, _Up>>
+    : _CUDA_VSTD::tuple_element<_Ip, const volatile _CUDA_VSTD::pair<_Tp, _Up>>
+{};
+
+template <class... _Tp>
+struct tuple_size<_CUDA_VSTD::tuple<_Tp...>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
+{};
+
+template <class... _Tp>
+struct tuple_size<const _CUDA_VSTD::tuple<_Tp...>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
+{};
+
+template <class... _Tp>
+struct tuple_size<volatile _CUDA_VSTD::tuple<_Tp...>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
+{};
+
+template <class... _Tp>
+struct tuple_size<const volatile _CUDA_VSTD::tuple<_Tp...>> : _CUDA_VSTD::tuple_size<_CUDA_VSTD::tuple<_Tp...>>
+{};
+
+template <size_t _Ip, class... _Tp>
+struct tuple_element<_Ip, _CUDA_VSTD::tuple<_Tp...>> : _CUDA_VSTD::tuple_element<_Ip, _CUDA_VSTD::tuple<_Tp...>>
+{};
+
+template <size_t _Ip, class... _Tp>
+struct tuple_element<_Ip, const _CUDA_VSTD::tuple<_Tp...>>
+    : _CUDA_VSTD::tuple_element<_Ip, const _CUDA_VSTD::tuple<_Tp...>>
+{};
+
+template <size_t _Ip, class... _Tp>
+struct tuple_element<_Ip, volatile _CUDA_VSTD::tuple<_Tp...>>
+    : _CUDA_VSTD::tuple_element<_Ip, volatile _CUDA_VSTD::tuple<_Tp...>>
+{};
+
+template <size_t _Ip, class... _Tp>
+struct tuple_element<_Ip, const volatile _CUDA_VSTD::tuple<_Tp...>>
+    : _CUDA_VSTD::tuple_element<_Ip, const volatile _CUDA_VSTD::tuple<_Tp...>>
+{};
+} // namespace std
+#  endif // _LIBCUDACXX_STD_VER > 14
+
+#  if defined(_LIBCUDACXX_COMPILER_CLANG)
+#    pragma clang diagnostic pop
+#  endif // _LIBCUDACXX_COMPILER_CLANG
 
 #endif // __cuda_std__
 
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_element.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_element.h
index 768f7e1edf..d2b3bdd4f6 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_element.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_element.h
@@ -11,7 +11,7 @@
 #define _LIBCUDACXX___TUPLE_TUPLE_ELEMENT_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__tuple_dir/tuple_indices.h"
@@ -22,58 +22,65 @@
 #include "../cstddef"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template <size_t _Ip, class _Tp> struct _LIBCUDACXX_TEMPLATE_VIS tuple_element;
+template <size_t _Ip, class _Tp>
+struct _LIBCUDACXX_TEMPLATE_VIS tuple_element;
+
+template <size_t _Ip, class... _Tp>
+using __tuple_element_t _LIBCUDACXX_NODEBUG_TYPE = typename tuple_element<_Ip, _Tp...>::type;
 
 template <size_t _Ip, class _Tp>
 struct _LIBCUDACXX_TEMPLATE_VIS tuple_element<_Ip, const _Tp>
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename add_const<typename tuple_element<_Ip, _Tp>::type>::type type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE typename add_const<__tuple_element_t<_Ip, _Tp>>::type type;
 };
 
 template <size_t _Ip, class _Tp>
 struct _LIBCUDACXX_TEMPLATE_VIS tuple_element<_Ip, volatile _Tp>
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename add_volatile<typename tuple_element<_Ip, _Tp>::type>::type type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE typename add_volatile<__tuple_element_t<_Ip, _Tp>>::type type;
 };
 
 template <size_t _Ip, class _Tp>
 struct _LIBCUDACXX_TEMPLATE_VIS tuple_element<_Ip, const volatile _Tp>
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename add_cv<typename tuple_element<_Ip, _Tp>::type>::type type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE typename add_cv<__tuple_element_t<_Ip, _Tp>>::type type;
 };
 
 #ifdef _LIBCUDACXX_COMPILER_MSVC
 
 namespace __indexer_detail {
 
-template <size_t _Idx, class ..._Types>
+template <size_t _Idx, class... _Types>
 struct _nth_of;
 
-template <class _Head, class ..._Tail>
-struct _nth_of<0, _Head, _Tail...> {
-    using type = _Head;
+template <class _Head, class... _Tail>
+struct _nth_of<0, _Head, _Tail...>
+{
+  using type = _Head;
 };
 
-template <size_t _Idx, class _Head, class ..._Tail>
-struct _nth_of<_Idx, _Head, _Tail...> {
-    using type = typename _nth_of<_Idx-1, _Tail...>::type;
+template <size_t _Idx, class _Head, class... _Tail>
+struct _nth_of<_Idx, _Head, _Tail...>
+{
+  using type = typename _nth_of<_Idx - 1, _Tail...>::type;
 };
 
-template <size_t _Idx, class ..._Types>
-struct nth_of {
-    static_assert(_Idx < sizeof...(_Types), "");
-    using _impl = _nth_of<_Idx, _Types...>;
-    using type = typename _impl::type;
+template <size_t _Idx, class... _Types>
+struct nth_of
+{
+  static_assert(_Idx < sizeof...(_Types), "");
+  using _impl = _nth_of<_Idx, _Types...>;
+  using type  = typename _impl::type;
 };
 
 } // namespace __indexer_detail
 
-template <size_t _Idx, class ..._Types>
+template <size_t _Idx, class... _Types>
 using __type_pack_element _LIBCUDACXX_NODEBUG_TYPE = typename __indexer_detail::nth_of<_Idx, _Types...>::type;
 
 #elif !__has_builtin(__type_pack_element)
@@ -81,41 +88,39 @@ using __type_pack_element _LIBCUDACXX_NODEBUG_TYPE = typename __indexer_detail::
 namespace __indexer_detail {
 
 template <size_t _Idx, class _Tp>
-struct __indexed { using type _LIBCUDACXX_NODEBUG_TYPE = _Tp; };
+struct __indexed
+{
+  using type _LIBCUDACXX_NODEBUG_TYPE = _Tp;
+};
 
-template <class _Types, class _Indexes> struct __indexer;
+template <class _Types, class _Indexes>
+struct __indexer;
 
-template <class ..._Types, size_t ..._Idx>
-struct __indexer<__tuple_types<_Types...>, __tuple_indices<_Idx...>>
-    : __indexed<_Idx, _Types>...
+template <class... _Types, size_t... _Idx>
+struct __indexer<__tuple_types<_Types...>, __tuple_indices<_Idx...>> : __indexed<_Idx, _Types>...
 {};
 
 template <size_t _Idx, class _Tp>
-_LIBCUDACXX_INLINE_VISIBILITY
-__indexed<_Idx, _Tp> __at_index(__indexed<_Idx, _Tp> const&);
+_LIBCUDACXX_INLINE_VISIBILITY __indexed<_Idx, _Tp> __at_index(__indexed<_Idx, _Tp> const&);
 
 } // namespace __indexer_detail
 
-template <size_t _Idx, class ..._Types>
-using __type_pack_element _LIBCUDACXX_NODEBUG_TYPE = typename decltype(
-    __indexer_detail::__at_index<_Idx>(
-        __indexer_detail::__indexer<
-            __tuple_types<_Types...>,
-            typename __make_tuple_indices<sizeof...(_Types)>::type
-        >{})
-  )::type;
+template <size_t _Idx, class... _Types>
+using __type_pack_element _LIBCUDACXX_NODEBUG_TYPE =
+    typename decltype(__indexer_detail::__at_index<_Idx>(__indexer_detail::__indexer< __tuple_types<_Types...>,
+        __make_tuple_indices_t<sizeof...(_Types)>>{}))::type;
 #endif
 
-template <size_t _Ip, class ..._Types>
+template <size_t _Ip, class... _Types>
 struct _LIBCUDACXX_TEMPLATE_VIS tuple_element<_Ip, __tuple_types<_Types...> >
 {
-    static_assert(_Ip < sizeof...(_Types), "tuple_element index out of range");
-    typedef _LIBCUDACXX_NODEBUG_TYPE __type_pack_element<_Ip, _Types...> type;
+  static_assert(_Ip < sizeof...(_Types), "tuple_element index out of range");
+  typedef _LIBCUDACXX_NODEBUG_TYPE __type_pack_element<_Ip, _Types...> type;
 };
 
 #if _LIBCUDACXX_STD_VER > 11
-template <size_t _Ip, class ..._Tp>
-using tuple_element_t _LIBCUDACXX_NODEBUG_TYPE = typename tuple_element <_Ip, _Tp...>::type;
+template <size_t _Ip, class... _Tp>
+using tuple_element_t _LIBCUDACXX_NODEBUG_TYPE = typename tuple_element<_Ip, _Tp...>::type;
 #endif
 
 _LIBCUDACXX_END_NAMESPACE_STD
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_indices.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_indices.h
index 1964641e40..67787321c9 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_indices.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_indices.h
@@ -11,14 +11,14 @@
 #define _LIBCUDACXX___TUPLE_MAKE_TUPLE_INDICES_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__utility/integer_sequence.h"
 #include "../cstddef"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
@@ -26,10 +26,13 @@ _LIBCUDACXX_BEGIN_NAMESPACE_STD
 template <size_t _Ep, size_t _Sp = 0>
 struct __make_tuple_indices
 {
-    static_assert(_Sp <= _Ep, "__make_tuple_indices input error");
-    typedef __make_indices_imp<_Ep, _Sp> type;
+  static_assert(_Sp <= _Ep, "__make_tuple_indices input error");
+  typedef __make_indices_imp<_Ep, _Sp> type;
 };
 
+template <size_t _Ep, size_t _Sp = 0>
+using __make_tuple_indices_t = typename __make_tuple_indices<_Ep, _Sp>::type;
+
 _LIBCUDACXX_END_NAMESPACE_STD
 
 #endif // _LIBCUDACXX___TUPLE_MAKE_TUPLE_INDICES_H
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_like.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_like.h
index bf251a858a..8557bd5616 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_like.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_like.h
@@ -11,7 +11,7 @@
 #define _LIBCUDACXX___TUPLE_TUPLE_LIKE_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__fwd/array.h"
@@ -22,24 +22,40 @@
 #include "../cstddef"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template <class _Tp> struct __tuple_like : false_type {};
-
-template <class _Tp> struct __tuple_like<const _Tp> : public __tuple_like<_Tp> {};
-template <class _Tp> struct __tuple_like<volatile _Tp> : public __tuple_like<_Tp> {};
-template <class _Tp> struct __tuple_like<const volatile _Tp> : public __tuple_like<_Tp> {};
-
-template <class... _Tp> struct __tuple_like<tuple<_Tp...> > : true_type {};
-
-template <class _T1, class _T2> struct __tuple_like<pair<_T1, _T2> > : true_type {};
-
-template <class _Tp, size_t _Size> struct __tuple_like<array<_Tp, _Size> > : true_type {};
-
-template <class... _Tp> struct __tuple_like<__tuple_types<_Tp...> > : true_type {};
+template <class _Tp>
+struct __tuple_like : false_type
+{};
+
+template <class _Tp>
+struct __tuple_like<const _Tp> : public __tuple_like<_Tp>
+{};
+template <class _Tp>
+struct __tuple_like<volatile _Tp> : public __tuple_like<_Tp>
+{};
+template <class _Tp>
+struct __tuple_like<const volatile _Tp> : public __tuple_like<_Tp>
+{};
+
+template <class... _Tp>
+struct __tuple_like<tuple<_Tp...> > : true_type
+{};
+
+template <class _T1, class _T2>
+struct __tuple_like<pair<_T1, _T2> > : true_type
+{};
+
+template <class _Tp, size_t _Size>
+struct __tuple_like<array<_Tp, _Size> > : true_type
+{};
+
+template <class... _Tp>
+struct __tuple_like<__tuple_types<_Tp...> > : true_type
+{};
 
 _LIBCUDACXX_END_NAMESPACE_STD
 
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_size.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_size.h
index 9a97ccce34..74c8ebd1bc 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_size.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_size.h
@@ -11,7 +11,7 @@
 #define _LIBCUDACXX___TUPLE_TUPLE_SIZE_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "../__fwd/tuple.h"
@@ -23,47 +23,47 @@
 #include "../cstddef"
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template <class _Tp> struct _LIBCUDACXX_TEMPLATE_VIS tuple_size;
+template <class _Tp>
+struct _LIBCUDACXX_TEMPLATE_VIS tuple_size;
 
 template <class _Tp, class...>
 using __enable_if_tuple_size_imp = _Tp;
 
 template <class _Tp>
-struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<__enable_if_tuple_size_imp<
-    const _Tp,
-    __enable_if_t<!is_volatile<_Tp>::value>,
-    integral_constant<size_t, sizeof(tuple_size<_Tp>)>>>
-    : public integral_constant<size_t, tuple_size<_Tp>::value> {};
+struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<__enable_if_tuple_size_imp< const _Tp,
+    __enable_if_t<!is_volatile<_Tp>::value>, integral_constant<size_t, sizeof(tuple_size<_Tp>)>>>
+    : public integral_constant<size_t, tuple_size<_Tp>::value>
+{};
 
 template <class _Tp>
-struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<__enable_if_tuple_size_imp<
-    volatile _Tp,
-    __enable_if_t<!is_const<_Tp>::value>,
-    integral_constant<size_t, sizeof(tuple_size<_Tp>)>>>
-    : public integral_constant<size_t, tuple_size<_Tp>::value> {};
+struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<__enable_if_tuple_size_imp< volatile _Tp,
+    __enable_if_t<!is_const<_Tp>::value>, integral_constant<size_t, sizeof(tuple_size<_Tp>)>>>
+    : public integral_constant<size_t, tuple_size<_Tp>::value>
+{};
 
 template <class _Tp>
-struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<__enable_if_tuple_size_imp<
-    const volatile _Tp,
-    integral_constant<size_t, sizeof(tuple_size<_Tp>)>>>
-    : public integral_constant<size_t, tuple_size<_Tp>::value> {};
+struct _LIBCUDACXX_TEMPLATE_VIS
+    tuple_size<__enable_if_tuple_size_imp< const volatile _Tp, integral_constant<size_t, sizeof(tuple_size<_Tp>)>>>
+    : public integral_constant<size_t, tuple_size<_Tp>::value>
+{};
+
+template <class... _Tp>
+struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<tuple<_Tp...> > : public integral_constant<size_t, sizeof...(_Tp)>
+{};
 
-template <class ..._Tp>
-struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<tuple<_Tp...> >
-    : public integral_constant<size_t, sizeof...(_Tp)>
-{
-};
+template <class... _Tp>
+struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<__tuple_types<_Tp...> > : public integral_constant<size_t, sizeof...(_Tp)>
+{};
 
-template <class ..._Tp>
-struct _LIBCUDACXX_TEMPLATE_VIS tuple_size<__tuple_types<_Tp...> >
-    : public integral_constant<size_t, sizeof...(_Tp)>
-{
-};
+#if _LIBCUDACXX_STD_VER >= 17
+template <class _Tp>
+_LIBCUDACXX_INLINE_VAR constexpr size_t tuple_size_v = tuple_size<_Tp>::value;
+#endif // _LIBCUDACXX_STD_VER >= 17
 
 _LIBCUDACXX_END_NAMESPACE_STD
 
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_types.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_types.h
index ecb770aab1..c98f7f23c3 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_types.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__tuple_dir/tuple_types.h
@@ -11,16 +11,18 @@
 #define _LIBCUDACXX___TUPLE_TUPLE_TYPES_H
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template <class ..._Tp> struct __tuple_types {};
+template <class... _Tp>
+struct __tuple_types
+{};
 
 _LIBCUDACXX_END_NAMESPACE_STD
 
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__utility/pair.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__utility/pair.h
index 4e336f746a..bc6869ee4e 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/__utility/pair.h
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/__utility/pair.h
@@ -366,8 +366,8 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair
         noexcept((is_nothrow_constructible<first_type, _Args1...>::value &&
                     is_nothrow_constructible<second_type, _Args2...>::value))
         : pair(__pc, __first_args, __second_args,
-                typename __make_tuple_indices<sizeof...(_Args1)>::type(),
-                typename __make_tuple_indices<sizeof...(_Args2) >::type()) {}
+                __make_tuple_indices_t<sizeof...(_Args1)>(),
+                __make_tuple_indices_t<sizeof...(_Args2)>()) {}
 
     _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX17
     pair& operator=(__conditional_t<
@@ -769,32 +769,28 @@ struct __get_pair<1>
 };
 
 template <size_t _Ip, class _T1, class _T2>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, pair<_T1, _T2> >::type&
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, pair<_T1, _T2>>&
 get(pair<_T1, _T2>& __p) noexcept
 {
-    return __get_pair<_Ip>::get(__p);
+  return __get_pair<_Ip>::get(__p);
 }
 
 template <size_t _Ip, class _T1, class _T2>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, pair<_T1, _T2> >::type&
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, pair<_T1, _T2>>&
 get(const pair<_T1, _T2>& __p) noexcept
 {
-    return __get_pair<_Ip>::get(__p);
+  return __get_pair<_Ip>::get(__p);
 }
 
 template <size_t _Ip, class _T1, class _T2>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, pair<_T1, _T2> >::type&&
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, pair<_T1, _T2>>&&
 get(pair<_T1, _T2>&& __p) noexcept
 {
     return __get_pair<_Ip>::get(_CUDA_VSTD::move(__p));
 }
 
 template <size_t _Ip, class _T1, class _T2>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, pair<_T1, _T2> >::type&&
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, pair<_T1, _T2>>&&
 get(const pair<_T1, _T2>&& __p) noexcept
 {
     return __get_pair<_Ip>::get(_CUDA_VSTD::move(__p));
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/tuple b/libcudacxx/include/cuda/std/detail/libcxx/include/tuple
index 6dac4ab7c4..566f827e89 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/tuple
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/tuple
@@ -146,14 +146,13 @@ template <class... Types>
 */
 
 #ifndef __cuda_std__
-#include <__config>
+#  include <__config>
 #endif // __cuda_std__
 
 #include "__assert" // all public C++ headers provide the assertion handler
-#include "__functional_base"
 #include "__functional/unwrap_ref.h"
+#include "__functional_base"
 #include "__fwd/array.h"
-#include "__type_traits/maybe_const.h"
 #include "__tuple_dir/apply_cv.h"
 #include "__tuple_dir/make_tuple_types.h"
 #include "__tuple_dir/sfinae_helpers.h"
@@ -163,6 +162,7 @@ template <class... Types>
 #include "__tuple_dir/tuple_like.h"
 #include "__tuple_dir/tuple_size.h"
 #include "__tuple_dir/tuple_types.h"
+#include "__type_traits/maybe_const.h"
 #include "__utility/forward.h"
 #include "__utility/integer_sequence.h"
 #include "__utility/move.h"
@@ -172,918 +172,779 @@ template <class... Types>
 #include "climits"
 #include "cstddef"
 #include "cstdint"
-#include "utility"
 #include "type_traits"
+#include "utility"
 
 // standard-mandated includes
 #include "version"
 
 // [tuple.syn]
 #ifndef _LIBCUDACXX_HAS_NO_SPACESHIP_OPERATOR
-#include "compare"
+#  include "compare"
 #endif
 
 #ifndef __cuda_std__
-#include <__pragma_push>
+#  include <__pragma_push>
 #endif // __cuda_std__
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
-#pragma GCC system_header
+#  pragma GCC system_header
 #endif
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-
 // __tuple_leaf
+struct __tuple_leaf_default_constructor_tag
+{};
 
-template <size_t _Ip, class _Hp,
-          bool=is_empty<_Hp>::value && !__libcpp_is_final<_Hp>::value
-         >
+template <size_t _Ip, class _Hp, bool = _LIBCUDACXX_TRAIT(is_empty, _Hp) && !__libcpp_is_final<_Hp>::value>
 class __tuple_leaf;
 
 template <size_t _Ip, class _Hp, bool _Ep>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-void swap(__tuple_leaf<_Ip, _Hp, _Ep>& __x, __tuple_leaf<_Ip, _Hp, _Ep>& __y)
-    noexcept(__is_nothrow_swappable<_Hp>::value)
+inline _LIBCUDACXX_INLINE_VISIBILITY void
+swap(__tuple_leaf<_Ip, _Hp, _Ep>& __x, __tuple_leaf<_Ip, _Hp, _Ep>& __y) noexcept(__is_nothrow_swappable<_Hp>::value)
 {
-    swap(__x.get(), __y.get());
+  swap(__x.get(), __y.get());
 }
 
 template <size_t _Ip, class _Hp, bool>
 class __tuple_leaf
 {
-    _Hp __value_;
+  _Hp __value_;
 
-    template <class _Tp>
-    _LIBCUDACXX_INLINE_VISIBILITY
-    static constexpr bool __can_bind_reference() {
+  template <class _Tp>
+  _LIBCUDACXX_INLINE_VISIBILITY static constexpr bool
+  __can_bind_reference()
+  {
 #if __has_keyword(__reference_binds_to_temporary)
-      return !__reference_binds_to_temporary(_Hp, _Tp);
+    return !__reference_binds_to_temporary(_Hp, _Tp);
 #else
-      return true;
+    return true;
 #endif
-    }
+  }
 
-    _LIBCUDACXX_INLINE_VISIBILITY __tuple_leaf& operator=(const __tuple_leaf&);
-public:
-    _LIBCUDACXX_INLINE_VISIBILITY constexpr __tuple_leaf()
-             noexcept(is_nothrow_default_constructible<_Hp>::value) : __value_()
-       {static_assert(!is_reference<_Hp>::value,
-              "Attempted to default construct a reference element in a tuple");}
-
-    template <class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf(integral_constant<int, 0>, const _Alloc&)
-            : __value_()
-        {static_assert(!is_reference<_Hp>::value,
-              "Attempted to default construct a reference element in a tuple");}
-
-    template <class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a)
-            : __value_(allocator_arg_t(), __a)
-        {static_assert(!is_reference<_Hp>::value,
-              "Attempted to default construct a reference element in a tuple");}
-
-    template <class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a)
-            : __value_(__a)
-        {static_assert(!is_reference<_Hp>::value,
-              "Attempted to default construct a reference element in a tuple");}
-
-    template<class _Tp>
-    using __can_forward = _And<_IsNotSame<__remove_cvref_t<_Tp>, __tuple_leaf>,
-                               is_constructible<_Hp, _Tp>>;
-
-    template <class _Tp,
-              __enable_if_t<__can_forward<_Tp>::value, int> = 0>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        explicit __tuple_leaf(_Tp&& __t) noexcept((is_nothrow_constructible<_Hp, _Tp>::value))
-            : __value_(_CUDA_VSTD::forward<_Tp>(__t))
-        {static_assert(__can_bind_reference<_Tp&&>(),
-       "Attempted construction of reference element binds to a temporary whose lifetime has ended");}
-
-    template <class _Tp, class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit __tuple_leaf(integral_constant<int, 0>, const _Alloc&, _Tp&& __t)
-            : __value_(_CUDA_VSTD::forward<_Tp>(__t))
-        {static_assert(__can_bind_reference<_Tp&&>(),
-       "Attempted construction of reference element binds to a temporary whose lifetime has ended");}
-
-    template <class _Tp, class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a, _Tp&& __t)
-            : __value_(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tp>(__t))
-        {static_assert(!is_reference<_Hp>::value,
-            "Attempted to uses-allocator construct a reference element in a tuple");}
-
-    template <class _Tp, class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a, _Tp&& __t)
-            : __value_(_CUDA_VSTD::forward<_Tp>(__t), __a)
-        {static_assert(!is_reference<_Hp>::value,
-           "Attempted to uses-allocator construct a reference element in a tuple");}
-
-    __tuple_leaf(const __tuple_leaf& __t) = default;
-    __tuple_leaf(__tuple_leaf&& __t) = default;
-
-    template <class _Tp>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf&
-        operator=(_Tp&& __t) noexcept((is_nothrow_assignable<_Hp&, _Tp>::value))
-        {
-            __value_ = _CUDA_VSTD::forward<_Tp>(__t);
-            return *this;
-        }
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    int swap(__tuple_leaf& __t) noexcept(__is_nothrow_swappable<__tuple_leaf>::value)
-    {
-        _CUDA_VSTD::swap(*this, __t);
-        return 0;
-    }
+  _LIBCUDACXX_INLINE_VISIBILITY __tuple_leaf& operator=(const __tuple_leaf&);
 
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11       _Hp& get()       noexcept {return __value_;}
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const _Hp& get() const noexcept {return __value_;}
+public:
+  _LIBCUDACXX_INLINE_VISIBILITY constexpr __tuple_leaf() noexcept(
+      _LIBCUDACXX_TRAIT(is_nothrow_default_constructible, _Hp))
+      : __value_()
+  {
+    static_assert(
+        !_LIBCUDACXX_TRAIT(is_reference, _Hp), "Attempted to default construct a reference element in a tuple");
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY constexpr __tuple_leaf(__tuple_leaf_default_constructor_tag) noexcept(
+      _LIBCUDACXX_TRAIT(is_nothrow_default_constructible, _Hp))
+      : __value_()
+  {
+    static_assert(
+        !_LIBCUDACXX_TRAIT(is_reference, _Hp), "Attempted to default construct a reference element in a tuple");
+  }
+
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  __tuple_leaf(integral_constant<int, 0>, const _Alloc&)
+      : __value_()
+  {
+    static_assert(
+        !_LIBCUDACXX_TRAIT(is_reference, _Hp), "Attempted to default construct a reference element in a tuple");
+  }
+
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a)
+      : __value_(allocator_arg_t(), __a)
+  {
+    static_assert(
+        !_LIBCUDACXX_TRAIT(is_reference, _Hp), "Attempted to default construct a reference element in a tuple");
+  }
+
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a)
+      : __value_(__a)
+  {
+    static_assert(
+        !_LIBCUDACXX_TRAIT(is_reference, _Hp), "Attempted to default construct a reference element in a tuple");
+  }
+
+  template <class _Tp>
+  using __can_forward = _And<_IsNotSame<__remove_cvref_t<_Tp>, __tuple_leaf>, is_constructible<_Hp, _Tp>>;
+
+  template <class _Tp, __enable_if_t<__can_forward<_Tp>::value, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit __tuple_leaf(_Tp&& __t) noexcept(
+      _LIBCUDACXX_TRAIT(is_nothrow_constructible, _Hp, _Tp))
+      : __value_(_CUDA_VSTD::forward<_Tp>(__t))
+  {
+    static_assert(__can_bind_reference<_Tp&&>(),
+        "Attempted construction of reference element binds to a "
+        "temporary whose lifetime has ended");
+  }
+
+  template <class _Tp, class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_leaf(integral_constant<int, 0>, const _Alloc&, _Tp&& __t)
+      : __value_(_CUDA_VSTD::forward<_Tp>(__t))
+  {
+    static_assert(__can_bind_reference<_Tp&&>(),
+        "Attempted construction of reference element binds to a "
+        "temporary whose lifetime has ended");
+  }
+
+  template <class _Tp, class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a, _Tp&& __t)
+      : __value_(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tp>(__t))
+  {
+    static_assert(
+        !_LIBCUDACXX_TRAIT(is_reference, _Hp), "Attempted to uses-allocator construct a reference element in a tuple");
+  }
+
+  template <class _Tp, class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a, _Tp&& __t)
+      : __value_(_CUDA_VSTD::forward<_Tp>(__t), __a)
+  {
+    static_assert(
+        !_LIBCUDACXX_TRAIT(is_reference, _Hp), "Attempted to uses-allocator construct a reference element in a tuple");
+  }
+
+  __tuple_leaf(const __tuple_leaf& __t) = default;
+  __tuple_leaf(__tuple_leaf&& __t)      = default;
+
+  template <class _Tp>
+  _LIBCUDACXX_INLINE_VISIBILITY __tuple_leaf&
+  operator=(_Tp&& __t) noexcept(_LIBCUDACXX_TRAIT(is_nothrow_assignable, _Hp&, _Tp))
+  {
+    __value_ = _CUDA_VSTD::forward<_Tp>(__t);
+    return *this;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY int
+  swap(__tuple_leaf& __t) noexcept(__is_nothrow_swappable<__tuple_leaf>::value)
+  {
+    _CUDA_VSTD::swap(*this, __t);
+    return 0;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _Hp&
+  get() noexcept
+  {
+    return __value_;
+  }
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const _Hp&
+  get() const noexcept
+  {
+    return __value_;
+  }
 };
 
 template <size_t _Ip, class _Hp>
-class __tuple_leaf<_Ip, _Hp, true>
-    : private _Hp
+class __tuple_leaf<_Ip, _Hp, true> : private _Hp
 {
+  _LIBCUDACXX_INLINE_VISIBILITY __tuple_leaf& operator=(const __tuple_leaf&);
 
-    _LIBCUDACXX_INLINE_VISIBILITY __tuple_leaf& operator=(const __tuple_leaf&);
 public:
-    _LIBCUDACXX_INLINE_VISIBILITY constexpr __tuple_leaf()
-             noexcept(is_nothrow_default_constructible<_Hp>::value) {}
-
-    template <class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf(integral_constant<int, 0>, const _Alloc&) {}
-
-    template <class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a)
-            : _Hp(allocator_arg_t(), __a) {}
-
-    template <class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a)
-            : _Hp(__a) {}
-
-    template<class _Tp>
-    using __can_forward = _And<_IsNotSame<__remove_cvref_t<_Tp>, __tuple_leaf>,
-                               is_constructible<_Hp, _Tp>>;
-
-    template <class _Tp,
-              __enable_if_t<__can_forward<_Tp>::value, int> = 0>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    explicit __tuple_leaf(_Tp&& __t) noexcept((is_nothrow_constructible<_Hp, _Tp>::value))
-        : _Hp(_CUDA_VSTD::forward<_Tp>(__t)) {}
-
-    template <class _Tp, class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit __tuple_leaf(integral_constant<int, 0>, const _Alloc&, _Tp&& __t)
-            : _Hp(_CUDA_VSTD::forward<_Tp>(__t)) {}
-
-    template <class _Tp, class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a, _Tp&& __t)
-            : _Hp(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tp>(__t)) {}
-
-    template <class _Tp, class _Alloc>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a, _Tp&& __t)
-            : _Hp(_CUDA_VSTD::forward<_Tp>(__t), __a) {}
-
-    __tuple_leaf(__tuple_leaf const &) = default;
-    __tuple_leaf(__tuple_leaf &&) = default;
-
-    template <class _Tp>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_leaf&
-        operator=(_Tp&& __t) noexcept((is_nothrow_assignable<_Hp&, _Tp>::value))
-        {
-            _Hp::operator=(_CUDA_VSTD::forward<_Tp>(__t));
-            return *this;
-        }
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    int
-    swap(__tuple_leaf& __t) noexcept(__is_nothrow_swappable<__tuple_leaf>::value)
-    {
-        _CUDA_VSTD::swap(*this, __t);
-        return 0;
-    }
-
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11       _Hp& get()       noexcept {return static_cast<_Hp&>(*this);}
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const _Hp& get() const noexcept {return static_cast<const _Hp&>(*this);}
+  _LIBCUDACXX_INLINE_VISIBILITY constexpr __tuple_leaf() noexcept(is_nothrow_default_constructible<_Hp>::value) {}
+
+  _LIBCUDACXX_INLINE_VISIBILITY constexpr __tuple_leaf(__tuple_leaf_default_constructor_tag) noexcept(
+      _LIBCUDACXX_TRAIT(is_nothrow_default_constructible, _Hp))
+      : _Hp()
+  {}
+
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  __tuple_leaf(integral_constant<int, 0>, const _Alloc&)
+  {}
+
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a)
+      : _Hp(allocator_arg_t(), __a)
+  {}
+
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a)
+      : _Hp(__a)
+  {}
+
+  template <class _Tp>
+  using __can_forward = _And<_IsNotSame<__remove_cvref_t<_Tp>, __tuple_leaf>, is_constructible<_Hp, _Tp>>;
+
+  template <class _Tp, __enable_if_t<__can_forward<_Tp>::value, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit __tuple_leaf(_Tp&& __t) noexcept(
+      (is_nothrow_constructible<_Hp, _Tp>::value))
+      : _Hp(_CUDA_VSTD::forward<_Tp>(__t))
+  {}
+
+  template <class _Tp, class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_leaf(integral_constant<int, 0>, const _Alloc&, _Tp&& __t)
+      : _Hp(_CUDA_VSTD::forward<_Tp>(__t))
+  {}
+
+  template <class _Tp, class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_leaf(integral_constant<int, 1>, const _Alloc& __a, _Tp&& __t)
+      : _Hp(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tp>(__t))
+  {}
+
+  template <class _Tp, class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_leaf(integral_constant<int, 2>, const _Alloc& __a, _Tp&& __t)
+      : _Hp(_CUDA_VSTD::forward<_Tp>(__t), __a)
+  {}
+
+  __tuple_leaf(__tuple_leaf const&) = default;
+  __tuple_leaf(__tuple_leaf&&)      = default;
+
+  template <class _Tp>
+  _LIBCUDACXX_INLINE_VISIBILITY __tuple_leaf&
+  operator=(_Tp&& __t) noexcept(_LIBCUDACXX_TRAIT(is_nothrow_assignable, _Hp&, _Tp))
+  {
+    _Hp::operator=(_CUDA_VSTD::forward<_Tp>(__t));
+    return *this;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY int
+  swap(__tuple_leaf& __t) noexcept(__is_nothrow_swappable<__tuple_leaf>::value)
+  {
+    _CUDA_VSTD::swap(*this, __t);
+    return 0;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _Hp&
+  get() noexcept
+  {
+    return static_cast<_Hp&>(*this);
+  }
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const _Hp&
+  get() const noexcept
+  {
+    return static_cast<const _Hp&>(*this);
+  }
 };
 
-template <class ..._Tp>
-_LIBCUDACXX_INLINE_VISIBILITY
-void __swallow(_Tp&&...) noexcept {}
+template <class... _Tp>
+_LIBCUDACXX_INLINE_VISIBILITY void
+__swallow(_Tp&&...) noexcept
+{}
 
 template <class _Tp>
 struct __all_default_constructible;
 
-template <class ..._Tp>
-struct __all_default_constructible<__tuple_types<_Tp...>>
-    : __all<is_default_constructible<_Tp>::value...>
-{ };
+template <class... _Tp>
+struct __all_default_constructible<__tuple_types<_Tp...>> : __all<_LIBCUDACXX_TRAIT(is_default_constructible, _Tp)...>
+{};
+
+struct __tuple_variadic_constructor_tag
+{};
 
 // __tuple_impl
 
-template<class _Indx, class ..._Tp> struct __tuple_impl;
+template <class _Indx, class... _Tp>
+struct __tuple_impl;
 
-template<size_t ..._Indx, class ..._Tp>
+template <size_t... _Indx, class... _Tp>
 struct _LIBCUDACXX_DECLSPEC_EMPTY_BASES __tuple_impl<__tuple_indices<_Indx...>, _Tp...>
     : public __tuple_leaf<_Indx, _Tp>...
 {
-    _LIBCUDACXX_INLINE_VISIBILITY
-    constexpr __tuple_impl()
-        noexcept(__all<is_nothrow_default_constructible<_Tp>::value...>::value) {}
-
-
-    // Handle non-allocator, full initialization
-    template <size_t ..._Uf, class ..._Tf,
-              size_t ..._Ul, class ..._Tl, class ..._Up,
-              _EnableIf<sizeof...(_Up) == sizeof...(_Tp), bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        explicit
-        __tuple_impl(__tuple_indices<_Uf...>, __tuple_types<_Tf...>,
-                     __tuple_indices<_Ul...>, __tuple_types<_Tl...>,
-                     _Up&&... __u)
-                     noexcept((__all<is_nothrow_constructible<_Tf, _Up>::value...>::value)) :
-            __tuple_leaf<_Uf, _Tf>(_CUDA_VSTD::forward<_Up>(__u))...
-            {}
-
-    // Handle non-allocator, partial default initialization
-    template <size_t ..._Uf, class ..._Tf,
-              size_t ..._Ul, class ..._Tl, class ..._Up,
-              _EnableIf<(sizeof...(_Up) > 0) && (sizeof...(_Up) < sizeof...(_Tp)), bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        explicit
-        __tuple_impl(__tuple_indices<_Uf...>, __tuple_types<_Tf...>,
-                     __tuple_indices<_Ul...>, __tuple_types<_Tl...>,
-                     _Up&&... __u)
-                     noexcept((__all<is_nothrow_constructible<_Tf, _Up>::value...>::value &&
-                                 __all<is_nothrow_default_constructible<_Tl>::value...>::value)) :
-            __tuple_leaf<_Uf, _Tf>(_CUDA_VSTD::forward<_Up>(__u))...,
-            __tuple_leaf<_Ul, _Tl>()...
-            {}
-
-    // Handle non-allocator, full default initialization
-    template <size_t ..._Uf, class ..._Tf,
-              size_t ..._Ul, class ..._Tl, class ..._Up,
-              _EnableIf<sizeof...(_Up) == 0, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        explicit
-        __tuple_impl(__tuple_indices<_Uf...>, __tuple_types<_Tf...>,
-                     __tuple_indices<_Ul...>, __tuple_types<_Tl...>,
-                     _Up&&...)
-                     noexcept((__all<is_nothrow_default_constructible<_Tl>::value...>::value)) :
-            __tuple_leaf<_Ul, _Tl>()...
-            {}
-
-
-    // Handle allocator aware, full initialization
-    template <class _Alloc, size_t ..._Uf, class ..._Tf,
-              size_t ..._Ul, class ..._Tl, class ..._Up,
-              _EnableIf<sizeof...(_Up) == sizeof...(_Tp), bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit
-        __tuple_impl(allocator_arg_t, const _Alloc& __a,
-                     __tuple_indices<_Uf...>, __tuple_types<_Tf...>,
-                     __tuple_indices<_Ul...>, __tuple_types<_Tl...>,
-                     _Up&&... __u) :
-            __tuple_leaf<_Uf, _Tf>(__uses_alloc_ctor<_Tf, _Alloc, _Up>(), __a,
-                    _CUDA_VSTD::forward<_Up>(__u))...
-            {}
-
-    // Handle allocator aware, partial default initialization
-    template <class _Alloc, size_t ..._Uf, class ..._Tf,
-              size_t ..._Ul, class ..._Tl, class ..._Up,
-              _EnableIf<(sizeof...(_Up) > 0) && (sizeof...(_Up) < sizeof...(_Tp)), bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit
-        __tuple_impl(allocator_arg_t, const _Alloc& __a,
-                     __tuple_indices<_Uf...>, __tuple_types<_Tf...>,
-                     __tuple_indices<_Ul...>, __tuple_types<_Tl...>,
-                     _Up&&... __u) :
-            __tuple_leaf<_Uf, _Tf>(__uses_alloc_ctor<_Tf, _Alloc, _Up>(), __a,
-                    _CUDA_VSTD::forward<_Up>(__u))...,
-            __tuple_leaf<_Ul, _Tl>(__uses_alloc_ctor<_Tl, _Alloc>(), __a)...
-            {}
-
-    // Handle allocator aware, full default initialization
-    template <class _Alloc, size_t ..._Uf, class ..._Tf,
-              size_t ..._Ul, class ..._Tl, class ..._Up,
-              _EnableIf<sizeof...(_Up) == 0, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit
-        __tuple_impl(allocator_arg_t, const _Alloc& __a,
-                     __tuple_indices<_Uf...>, __tuple_types<_Tf...>,
-                     __tuple_indices<_Ul...>, __tuple_types<_Tl...>,
-                     _Up&&...) :
-            __tuple_leaf<_Ul, _Tl>(__uses_alloc_ctor<_Tl, _Alloc>(), __a)...
-            {}
-
-    template <class _Tuple,
-              _EnableIf<
-                    __tuple_constructible<_Tuple, tuple<_Tp...> >::value,
-                    bool
-                > = false
-             >
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        __tuple_impl(_Tuple&& __t) noexcept((__all<is_nothrow_constructible<_Tp, typename tuple_element<_Indx,
-                                       typename __make_tuple_types<_Tuple>::type>::type>::value...>::value))
-            : __tuple_leaf<_Indx, _Tp>(_CUDA_VSTD::forward<typename tuple_element<_Indx,
-                                       typename __make_tuple_types<_Tuple>::type>::type>(_CUDA_VSTD::get<_Indx>(__t)))...
-            {}
-
-    template <class _Alloc, class _Tuple,
-              _EnableIf<
-                    __tuple_constructible<_Tuple, tuple<_Tp...> >::value,
-                    bool
-                > = false
-             >
-        _LIBCUDACXX_INLINE_VISIBILITY
-        __tuple_impl(allocator_arg_t, const _Alloc& __a, _Tuple&& __t)
-            : __tuple_leaf<_Indx, _Tp>(__uses_alloc_ctor<_Tp, _Alloc, typename tuple_element<_Indx,
-                                       typename __make_tuple_types<_Tuple>::type>::type>(), __a,
-                                       _CUDA_VSTD::forward<typename tuple_element<_Indx,
-                                       typename __make_tuple_types<_Tuple>::type>::type>(_CUDA_VSTD::get<_Indx>(__t)))...
-            {}
-
-    template <class _Tuple>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        typename enable_if
-        <
-            __tuple_assignable<_Tuple, tuple<_Tp...> >::value,
-            __tuple_impl&
-        >::type
-        operator=(_Tuple&& __t) noexcept((__all<is_nothrow_assignable<_Tp&, typename tuple_element<_Indx,
-                                       typename __make_tuple_types<_Tuple>::type>::type>::value...>::value))
-        {
-            __swallow(__tuple_leaf<_Indx, _Tp>::operator=(_CUDA_VSTD::forward<typename tuple_element<_Indx,
-                                       typename __make_tuple_types<_Tuple>::type>::type>(_CUDA_VSTD::get<_Indx>(__t)))...);
-            return *this;
-        }
-
-    __tuple_impl(const __tuple_impl&) = default;
-    __tuple_impl(__tuple_impl&&) = default;
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    __tuple_impl&
-    operator=(const __tuple_impl& __t) noexcept((__all<is_nothrow_copy_assignable<_Tp>::value...>::value))
-    {
-        __swallow(__tuple_leaf<_Indx, _Tp>::operator=(static_cast<const __tuple_leaf<_Indx, _Tp>&>(__t).get())...);
-        return *this;
-    }
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    __tuple_impl&
-    operator=(__tuple_impl&& __t) noexcept((__all<is_nothrow_move_assignable<_Tp>::value...>::value))
-    {
-        __swallow(__tuple_leaf<_Indx, _Tp>::operator=(_CUDA_VSTD::forward<_Tp>(static_cast<__tuple_leaf<_Indx, _Tp>&>(__t).get()))...);
-        return *this;
-    }
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    void swap(__tuple_impl& __t)
-        noexcept(__all<__is_nothrow_swappable<_Tp>::value...>::value)
-    {
-        __swallow(__tuple_leaf<_Indx, _Tp>::swap(static_cast<__tuple_leaf<_Indx, _Tp>&>(__t))...);
-    }
+  _LIBCUDACXX_INLINE_VISIBILITY constexpr __tuple_impl() noexcept(
+      __all<_LIBCUDACXX_TRAIT(is_nothrow_default_constructible, _Tp)...>::value)
+  {}
+
+  // Handle non-allocator, full initialization
+  // Old MSVC cannot handle the noexept specifier outside of template arguments
+  template <class... _Up, __enable_if_t<sizeof...(_Up) == sizeof...(_Tp), int> = 0,
+      bool __all_nothrow_constructible = __all<_LIBCUDACXX_TRAIT(is_nothrow_constructible, _Tp, _Up)...>::value>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit __tuple_impl(
+      __tuple_variadic_constructor_tag, _Up&&... __u) noexcept(__all_nothrow_constructible)
+      : __tuple_leaf<_Indx, _Tp>(_CUDA_VSTD::forward<_Up>(__u))...
+  {}
+
+  // Handle non-allocator, partial default initialization
+  // Recursively delegate until we have full rank
+  template <class... _Up, __enable_if_t<sizeof...(_Up) < sizeof...(_Tp), int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit __tuple_impl(
+      __tuple_variadic_constructor_tag __tag, _Up&&... __u) noexcept(noexcept(__tuple_impl(__tag,
+      _CUDA_VSTD::forward<_Up>(__u)..., __tuple_leaf_default_constructor_tag{})))
+      : __tuple_impl(__tag, _CUDA_VSTD::forward<_Up>(__u)..., __tuple_leaf_default_constructor_tag{})
+  {}
+
+  // Handle allocator aware, full initialization
+  template <class _Alloc, class... _Up, __enable_if_t<sizeof...(_Up) == sizeof...(_Tp), int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_impl(
+      allocator_arg_t, const _Alloc& __a, __tuple_variadic_constructor_tag, _Up&&... __u)
+      : __tuple_leaf<_Indx, _Tp>(__uses_alloc_ctor<_Tp, _Alloc, _Up>(), __a, _CUDA_VSTD::forward<_Up>(__u))...
+  {}
+
+  // Handle allocator aware, full default initialization
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit __tuple_impl(allocator_arg_t, const _Alloc& __a)
+      : __tuple_leaf<_Indx, _Tp>(__uses_alloc_ctor<_Tp, _Alloc>(), __a)...
+  {}
+
+  template <class _Tuple, size_t _Indx2>
+  using __tuple_elem_at = __tuple_element_t<_Indx2, __make_tuple_types_t<_Tuple>>;
+
+  template <class _Tuple, __enable_if_t<__tuple_constructible<_Tuple, tuple<_Tp...>>::value, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
+  __tuple_impl(_Tuple&& __t) noexcept(
+      (__all<_LIBCUDACXX_TRAIT(is_nothrow_constructible, _Tp, __tuple_elem_at<_Tuple, _Indx>)...>::value))
+      : __tuple_leaf<_Indx, _Tp>(_CUDA_VSTD::forward<__tuple_elem_at<_Tuple, _Indx>>(_CUDA_VSTD::get<_Indx>(__t)))...
+  {}
+
+  template <class _Alloc, class _Tuple, __enable_if_t<__tuple_constructible<_Tuple, tuple<_Tp...>>::value, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  __tuple_impl(allocator_arg_t, const _Alloc& __a, _Tuple&& __t)
+      : __tuple_leaf<_Indx, _Tp>(__uses_alloc_ctor<_Tp, _Alloc, __tuple_elem_at<_Tuple, _Indx>>(), __a,
+          _CUDA_VSTD::forward<__tuple_elem_at<_Tuple, _Indx>>(_CUDA_VSTD::get<_Indx>(__t)))...
+  {}
+
+  template < class _Tuple, __enable_if_t<__tuple_assignable<_Tuple, tuple<_Tp...>>::value, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY __tuple_impl&
+  operator=(_Tuple&& __t) noexcept(
+      (__all<_LIBCUDACXX_TRAIT(is_nothrow_assignable, _Tp&, __tuple_elem_at<_Tuple, _Indx>)...>::value))
+  {
+    __swallow(__tuple_leaf<_Indx, _Tp>::operator=(
+        _CUDA_VSTD::forward<__tuple_elem_at<_Tuple, _Indx>>(_CUDA_VSTD::get<_Indx>(__t)))...);
+    return *this;
+  }
+
+  __tuple_impl(const __tuple_impl&) = default;
+  __tuple_impl(__tuple_impl&&)      = default;
+
+  _LIBCUDACXX_INLINE_VISIBILITY __tuple_impl&
+  operator=(const __tuple_impl& __t) noexcept((__all<_LIBCUDACXX_TRAIT(is_nothrow_copy_assignable, _Tp)...>::value))
+  {
+    __swallow(__tuple_leaf<_Indx, _Tp>::operator=(static_cast<const __tuple_leaf<_Indx, _Tp>&>(__t).get())...);
+    return *this;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY __tuple_impl&
+  operator=(__tuple_impl&& __t) noexcept((__all<_LIBCUDACXX_TRAIT(is_nothrow_move_assignable, _Tp)...>::value))
+  {
+    __swallow(__tuple_leaf<_Indx, _Tp>::operator=(
+        _CUDA_VSTD::forward<_Tp>(static_cast<__tuple_leaf<_Indx, _Tp>&>(__t).get()))...);
+    return *this;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY void
+  swap(__tuple_impl& __t) noexcept(__all<__is_nothrow_swappable<_Tp>::value...>::value)
+  {
+    __swallow(__tuple_leaf<_Indx, _Tp>::swap(static_cast<__tuple_leaf<_Indx, _Tp>&>(__t))...);
+  }
 };
 
+struct __invalid_tuple_constraints
+{
+  static constexpr bool __implicit_constructible = false;
+  static constexpr bool __explicit_constructible = false;
+  static constexpr bool __nothrow_constructible  = false;
+};
 
+template <class... _Tp>
+struct __tuple_constraints
+{
+  static constexpr bool __implicit_default_constructible =
+      __all<__is_implicitly_default_constructible<_Tp>::value...>::value;
+
+  static constexpr bool __explicit_default_constructible =
+      !__implicit_default_constructible && __all<_LIBCUDACXX_TRAIT(is_default_constructible, _Tp)...>::value;
+
+  static constexpr bool __nothrow_default_constructible =
+      __all<_LIBCUDACXX_TRAIT(is_nothrow_default_constructible, _Tp)...>::value;
+
+  static constexpr bool __implicit_variadic_copy_constructible =
+      __tuple_constructible<tuple<const _Tp&...>, tuple<_Tp...>>::value
+      && __tuple_convertible<tuple<const _Tp&...>, tuple<_Tp...>>::value;
+
+  static constexpr bool __explicit_variadic_copy_constructible =
+      __tuple_constructible<tuple<const _Tp&...>, tuple<_Tp...>>::value
+      && !__tuple_convertible<tuple<const _Tp&...>, tuple<_Tp...>>::value;
+
+  static constexpr bool __nothrow_variadic_copy_constructible =
+      __all<_LIBCUDACXX_TRAIT(is_nothrow_copy_constructible, _Tp)...>::value;
+
+  template <class... _Args>
+  struct _PackExpandsToThisTuple : false_type
+  {};
+
+  template <class _Arg>
+  struct _PackExpandsToThisTuple<_Arg> : is_same<__remove_cvref_t<_Arg>, tuple<_Tp...>>
+  {};
+
+  template <class... _Args>
+  struct __variadic_constraints
+  {
+    static constexpr bool __implicit_constructible =
+        __tuple_constructible<tuple<_Args...>, tuple<_Tp...>>::value
+        && __tuple_convertible<tuple<_Args...>, tuple<_Tp...>>::value;
+
+    static constexpr bool __explicit_constructible =
+        __tuple_constructible<tuple<_Args...>, tuple<_Tp...>>::value
+        && !__tuple_convertible<tuple<_Args...>, tuple<_Tp...>>::value;
+
+    static constexpr bool __nothrow_constructible =
+        __all<_LIBCUDACXX_TRAIT(is_nothrow_constructible, _Tp, _Args)...>::value;
+  };
+
+  template <class... _Args>
+  struct __variadic_constraints_less_rank
+  {
+    static constexpr bool __implicit_constructible =
+        __tuple_constructible< tuple<_Args...>, __make_tuple_types_t<tuple<_Tp...>, sizeof...(_Args)>>::value
+        && __tuple_convertible< tuple<_Args...>, __make_tuple_types_t<tuple<_Tp...>, sizeof...(_Args)>>::value
+        && __all_default_constructible<__make_tuple_types_t< tuple<_Tp...>, sizeof...(_Tp), sizeof...(_Args)>>::value;
+
+    static constexpr bool __explicit_constructible =
+        __tuple_constructible< tuple<_Args...>, __make_tuple_types_t<tuple<_Tp...>, sizeof...(_Args)>>::value
+        && !__tuple_convertible< tuple<_Args...>, __make_tuple_types_t<tuple<_Tp...>, sizeof...(_Args)>>::value
+        && __all_default_constructible<__make_tuple_types_t< tuple<_Tp...>, sizeof...(_Tp), sizeof...(_Args)>>::value;
+  };
+
+  template <class _Tuple>
+  struct __valid_tuple_like_constraints
+  {
+    static constexpr bool __implicit_constructible =
+        __tuple_constructible<_Tuple, tuple<_Tp...>>::value && __tuple_convertible<_Tuple, tuple<_Tp...>>::value;
+
+    static constexpr bool __explicit_constructible =
+        __tuple_constructible<_Tuple, tuple<_Tp...>>::value && !__tuple_convertible<_Tuple, tuple<_Tp...>>::value;
+  };
+
+  template <class _Tuple>
+  struct __valid_tuple_like_constraints_rank_one
+  {
+    template <class _Tuple2>
+    struct _PreferTupleLikeConstructorImpl
+        : _Or<
+              // Don't attempt the two checks below if the tuple we are given
+              // has the same type as this tuple.
+              _IsSame<__remove_cvref_t<_Tuple2>, tuple<_Tp...>>,
+              _Lazy<_And, _Not<is_constructible<_Tp..., _Tuple2>>, _Not<is_convertible<_Tuple2, _Tp...>>>>
+    {};
+
+    // This trait is used to disable the tuple-like constructor when
+    // the UTypes... constructor should be selected instead.
+    // See LWG issue #2549.
+    template <class _Tuple2>
+    using _PreferTupleLikeConstructor = _PreferTupleLikeConstructorImpl<_Tuple2>;
+
+    static constexpr bool __implicit_constructible =
+        __tuple_constructible<_Tuple, tuple<_Tp...>>::value && __tuple_convertible<_Tuple, tuple<_Tp...>>::value
+        && _PreferTupleLikeConstructor<_Tuple>::value;
+
+    static constexpr bool __explicit_constructible =
+        __tuple_constructible<_Tuple, tuple<_Tp...>>::value && !__tuple_convertible<_Tuple, tuple<_Tp...>>::value
+        && _PreferTupleLikeConstructor<_Tuple>::value;
+  };
+
+  template <class _Tuple>
+  using __tuple_like_constraints =
+      _If<sizeof...(_Tp) == 1, __valid_tuple_like_constraints_rank_one<_Tuple>, __valid_tuple_like_constraints<_Tuple>>;
+};
 
-template <class ..._Tp>
+template <class... _Tp>
 class _LIBCUDACXX_TEMPLATE_VIS tuple
 {
-    typedef __tuple_impl<typename __make_tuple_indices<sizeof...(_Tp)>::type, _Tp...> _BaseT;
-
-    _BaseT __base_;
-
-#if defined(_LIBCUDACXX_ENABLE_TUPLE_IMPLICIT_REDUCED_ARITY_EXTENSION)
-    static constexpr bool _EnableImplicitReducedArityExtension = true;
-#else
-    static constexpr bool _EnableImplicitReducedArityExtension = false;
-#endif
+  typedef __tuple_impl<__make_tuple_indices_t<sizeof...(_Tp)>, _Tp...> _BaseT;
+
+  _BaseT __base_;
+
+  template <class... _Args>
+  struct _PackExpandsToThisTuple : false_type
+  {};
+
+  template <class _Arg>
+  struct _PackExpandsToThisTuple<_Arg> : is_same<__remove_cvref_t<_Arg>, tuple>
+  {};
+
+  template <size_t _Jp, class... _Up>
+  friend _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY __tuple_element_t<_Jp, tuple<_Up...>>& get(
+      tuple<_Up...>&) noexcept;
+  template <size_t _Jp, class... _Up>
+  friend _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY const __tuple_element_t<_Jp, tuple<_Up...>>&
+  get(const tuple<_Up...>&) noexcept;
+  template <size_t _Jp, class... _Up>
+  friend _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY __tuple_element_t<_Jp, tuple<_Up...>>&& get(
+      tuple<_Up...>&&) noexcept;
+  template <size_t _Jp, class... _Up>
+  friend _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY const __tuple_element_t<_Jp, tuple<_Up...>>&&
+  get(const tuple<_Up...>&&) noexcept;
 
-    template <class ..._Args>
-    struct _PackExpandsToThisTuple : false_type {};
-
-    template <class _Arg>
-    struct _PackExpandsToThisTuple<_Arg>
-        : is_same<__remove_cvref_t<_Arg>, tuple> {};
-
-    template <bool _MaybeEnable, class _Dummy = void>
-    struct _CheckArgsConstructor : __check_tuple_constructor_fail {};
-
-    template <class _Dummy>
-    struct _CheckArgsConstructor<true, _Dummy>
-    {
-        template <int&...>
-        struct __enable_implicit_default {
-            static constexpr bool value =
-                __all<__is_implicitly_default_constructible<_Tp>::value...>::value;
-        };
-
-        template <int&...>
-        struct __enable_explicit_default {
-            static constexpr bool value =
-                __all<is_default_constructible<_Tp>::value ...>::value &&
-                !__enable_implicit_default<>::value;
-        };
-
-        template <class ..._Args>
-        struct __enable_explicit {
-            static constexpr bool value =
-                __tuple_constructible<
-                    tuple<_Args...>,
-                    typename __make_tuple_types<tuple,
-                             sizeof...(_Args) < sizeof...(_Tp) ?
-                                 sizeof...(_Args) :
-                                 sizeof...(_Tp)>::type
-                >::value &&
-                !__tuple_convertible<
-                    tuple<_Args...>,
-                    typename __make_tuple_types<tuple,
-                             sizeof...(_Args) < sizeof...(_Tp) ?
-                                 sizeof...(_Args) :
-                                 sizeof...(_Tp)>::type
-                >::value &&
-                __all_default_constructible<
-                    typename __make_tuple_types<tuple, sizeof...(_Tp),
-                             sizeof...(_Args) < sizeof...(_Tp) ?
-                                 sizeof...(_Args) :
-                                 sizeof...(_Tp)>::type
-                >::value;
-        };
-
-        template <class ..._Args>
-        struct __enable_implicit {
-            static constexpr bool value =
-               __tuple_constructible<
-                    tuple<_Args...>,
-                    typename __make_tuple_types<tuple,
-                             sizeof...(_Args) < sizeof...(_Tp) ?
-                                 sizeof...(_Args) :
-                                 sizeof...(_Tp)>::type
-                >::value &&
-                __tuple_convertible<
-                    tuple<_Args...>,
-                    typename __make_tuple_types<tuple,
-                             sizeof...(_Args) < sizeof...(_Tp) ?
-                                 sizeof...(_Args) :
-                                 sizeof...(_Tp)>::type
-                >::value &&
-                __all_default_constructible<
-                    typename __make_tuple_types<tuple, sizeof...(_Tp),
-                             sizeof...(_Args) < sizeof...(_Tp) ?
-                                 sizeof...(_Args) :
-                                 sizeof...(_Tp)>::type
-                >::value;
-        };
-    };
-
-    template <bool _MaybeEnable,
-              bool = sizeof...(_Tp) == 1,
-              class _Dummy = void>
-    struct _CheckTupleLikeConstructor : __check_tuple_constructor_fail {};
-
-    template <class _Dummy>
-    struct _CheckTupleLikeConstructor<true, false, _Dummy>
-    {
-        template <class _Tuple>
-        struct __enable_implicit {
-            static constexpr bool value =
-                   __tuple_constructible<_Tuple, tuple>::value
-                && __tuple_convertible<_Tuple, tuple>::value;
-        };
-
-        template <class _Tuple>
-        struct __enable_explicit {
-            static constexpr bool value =
-                   __tuple_constructible<_Tuple, tuple>::value
-               && !__tuple_convertible<_Tuple, tuple>::value;
-        };
-    };
-
-    template <class _Dummy>
-    struct _CheckTupleLikeConstructor<true, true, _Dummy>
-    {
-        template <class _Tuple>
-        struct _PreferTupleLikeConstructorImpl :
-            _Or<
-                // Don't attempt the two checks below if the tuple we are given
-                // has the same type as this tuple.
-                _IsSame<__remove_cvref_t<_Tuple>, tuple>,
-                _Lazy<_And,
-                    _Not<is_constructible<_Tp..., _Tuple>>,
-                    _Not<is_convertible<_Tuple, _Tp...>>
-                >
-            >
-        { };
-
-        // This trait is used to disable the tuple-like constructor when
-        // the UTypes... constructor should be selected instead.
-        // See LWG issue #2549.
-        template <class _Tuple>
-        using _PreferTupleLikeConstructor = _PreferTupleLikeConstructorImpl<_Tuple>;
-
-        template <class _Tuple>
-        struct __enable_implicit {
-            static constexpr bool value =
-                _And<
-                    __tuple_constructible<_Tuple, tuple>,
-                    __tuple_convertible<_Tuple, tuple>,
-                    _PreferTupleLikeConstructor<_Tuple>
-                >::value;
-        };
-
-        template <class _Tuple>
-        struct __enable_explicit {
-            static constexpr bool value =
-                _And<
-                    __tuple_constructible<_Tuple, tuple>,
-                    _PreferTupleLikeConstructor<_Tuple>,
-                    _Not<__tuple_convertible<_Tuple, tuple>>
-                >::value;
-        };
-    };
-
-    template <size_t _Jp, class ..._Up> friend
-        _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY
-        typename tuple_element<_Jp, tuple<_Up...> >::type& get(tuple<_Up...>&) noexcept;
-    template <size_t _Jp, class ..._Up> friend
-        _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY
-        const typename tuple_element<_Jp, tuple<_Up...> >::type& get(const tuple<_Up...>&) noexcept;
-    template <size_t _Jp, class ..._Up> friend
-        _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY
-        typename tuple_element<_Jp, tuple<_Up...> >::type&& get(tuple<_Up...>&&) noexcept;
-    template <size_t _Jp, class ..._Up> friend
-        _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 _LIBCUDACXX_INLINE_VISIBILITY
-        const typename tuple_element<_Jp, tuple<_Up...> >::type&& get(const tuple<_Up...>&&) noexcept;
 public:
-
-    template<bool _Dummy>
-    struct _EnableConstructor {
-        static constexpr bool __implicit_default =
-            _CheckArgsConstructor<_Dummy>::template __enable_implicit_default<>::value;
-        static constexpr bool __explicit_default =
-            _CheckArgsConstructor<_Dummy>::template __enable_explicit_default<>::value;
-
-        static constexpr bool __implicit_variadic =
-            _CheckArgsConstructor<_Dummy>::template __enable_implicit<_Tp const&...>::value;
-        static constexpr bool __explicit_variadic =
-            _CheckArgsConstructor<_Dummy>::template __enable_explicit<_Tp const&...>::value;
-    };
-
-    template <bool _Dummy = true, __enable_if_t<_EnableConstructor<_Dummy>::__implicit_default, bool> = false>
-    _LIBCUDACXX_INLINE_VISIBILITY constexpr
-    tuple() noexcept(__all<is_nothrow_default_constructible<_Tp>::value...>::value) {}
-
-    template <bool _Dummy = true, __enable_if_t<_EnableConstructor<_Dummy>::__explicit_default, bool> = false>
-    explicit _LIBCUDACXX_INLINE_VISIBILITY constexpr
-    tuple() noexcept(__all<is_nothrow_default_constructible<_Tp>::value...>::value) {}
-
-    tuple(tuple const&) = default;
-    tuple(tuple&&) = default;
-
-    template <class _AllocArgT>
-    struct _EnableAllocConstructor {
-        static constexpr bool __implicit =
-            _CheckArgsConstructor<_IsSame<allocator_arg_t, _AllocArgT>::value >::template __enable_implicit_default<>::value;
-        static constexpr bool __explicit =
-            _CheckArgsConstructor<_IsSame<allocator_arg_t, _AllocArgT>::value >::template __enable_explicit_default<>::value;
-    };
-
-    template <class _AllocArgT, class _Alloc,
-              __enable_if_t<_EnableAllocConstructor<_AllocArgT>::__implicit, bool> = false>
-    _LIBCUDACXX_INLINE_VISIBILITY
-    tuple(_AllocArgT, _Alloc const& __a)
-      : __base_(allocator_arg_t(), __a,
-                    __tuple_indices<>(), __tuple_types<>(),
-                    typename __make_tuple_indices<sizeof...(_Tp), 0>::type(),
-                    __tuple_types<_Tp...>()) {}
-
-    template <class _AllocArgT, class _Alloc,
-              __enable_if_t<_EnableAllocConstructor<_AllocArgT>::__explicit, bool> = false>
-    explicit _LIBCUDACXX_INLINE_VISIBILITY
-    tuple(_AllocArgT, _Alloc const& __a)
-      : __base_(allocator_arg_t(), __a,
-                    __tuple_indices<>(), __tuple_types<>(),
-                    typename __make_tuple_indices<sizeof...(_Tp), 0>::type(),
-                    __tuple_types<_Tp...>()) {}
-
-    template <bool _Dummy = true, __enable_if_t<_EnableConstructor<_Dummy>::__implicit_variadic, bool> = false>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    tuple(const _Tp& ... __t) noexcept((__all<is_nothrow_copy_constructible<_Tp>::value...>::value))
-        : __base_(typename __make_tuple_indices<sizeof...(_Tp)>::type(),
-                typename __make_tuple_types<tuple, sizeof...(_Tp)>::type(),
-                typename __make_tuple_indices<0>::type(),
-                typename __make_tuple_types<tuple, 0>::type(),
-                __t...
-               ) {}
-
-    template <bool _Dummy = true, __enable_if_t<_EnableConstructor<_Dummy>::__explicit_variadic, bool> = false>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    explicit tuple(const _Tp& ... __t) noexcept((__all<is_nothrow_copy_constructible<_Tp>::value...>::value))
-        : __base_(typename __make_tuple_indices<sizeof...(_Tp)>::type(),
-                typename __make_tuple_types<tuple, sizeof...(_Tp)>::type(),
-                typename __make_tuple_indices<0>::type(),
-                typename __make_tuple_types<tuple, 0>::type(),
-                __t...
-               ) {}
-
-    template <class _Alloc, bool _Dummy = true, __enable_if_t<_EnableConstructor<_Dummy>::__implicit_variadic, bool> = false>
-      _LIBCUDACXX_INLINE_VISIBILITY
-      tuple(allocator_arg_t, const _Alloc& __a, const _Tp& ... __t)
-        : __base_(allocator_arg_t(), __a,
-                typename __make_tuple_indices<sizeof...(_Tp)>::type(),
-                typename __make_tuple_types<tuple, sizeof...(_Tp)>::type(),
-                typename __make_tuple_indices<0>::type(),
-                typename __make_tuple_types<tuple, 0>::type(),
-                __t...
-               ) {}
-
-    template <class _Alloc, bool _Dummy = true, __enable_if_t<_EnableConstructor<_Dummy>::__explicit_variadic, bool> = false>
-      _LIBCUDACXX_INLINE_VISIBILITY
-      explicit
-      tuple(allocator_arg_t, const _Alloc& __a, const _Tp& ... __t)
-        : __base_(allocator_arg_t(), __a,
-                typename __make_tuple_indices<sizeof...(_Tp)>::type(),
-                typename __make_tuple_types<tuple, sizeof...(_Tp)>::type(),
-                typename __make_tuple_indices<0>::type(),
-                typename __make_tuple_types<tuple, 0>::type(),
-                __t...
-               ) {}
+  template < class _Constraints                                          = __tuple_constraints<_Tp...>,
+      __enable_if_t<_Constraints::__implicit_default_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY constexpr tuple() noexcept(_Constraints::__nothrow_default_constructible)
+  {}
+
+  template < class _Constraints                                          = __tuple_constraints<_Tp...>,
+      __enable_if_t<_Constraints::__explicit_default_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit constexpr tuple() noexcept(_Constraints::__nothrow_default_constructible)
+  {}
+
+  tuple(tuple const&) = default;
+  tuple(tuple&&)      = default;
+
+  template <class _AllocArgT, class _Alloc, class _Constraints = __tuple_constraints<_Tp...>,
+      __enable_if_t<_LIBCUDACXX_TRAIT(is_same, allocator_arg_t, _AllocArgT), int> = 0,
+      __enable_if_t<_Constraints::__implicit_default_constructible, int>          = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(_AllocArgT, _Alloc const& __a) noexcept(_Constraints::__nothrow_default_constructible)
+      : __base_(allocator_arg_t(), __a)
+  {}
+
+  template <class _AllocArgT, class _Alloc, class _Constraints = __tuple_constraints<_Tp...>,
+      __enable_if_t<_LIBCUDACXX_TRAIT(is_same, allocator_arg_t, _AllocArgT), int> = 0,
+      __enable_if_t<_Constraints::__explicit_default_constructible, int>          = 0>
+  explicit _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(_AllocArgT, _Alloc const& __a) noexcept(_Constraints::__nothrow_default_constructible)
+      : __base_(allocator_arg_t(), __a)
+  {}
+
+  template <class _Constraints                                                 = __tuple_constraints<_Tp...>,
+      __enable_if_t<_Constraints::__implicit_variadic_copy_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
+  tuple(const _Tp&... __t) noexcept(_Constraints::__nothrow_variadic_copy_constructible)
+      : __base_(__tuple_variadic_constructor_tag{}, __t...)
+  {}
+
+  template <class _Constraints                                                 = __tuple_constraints<_Tp...>,
+      __enable_if_t<_Constraints::__explicit_variadic_copy_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit tuple(const _Tp&... __t) noexcept(
+      _Constraints::__nothrow_variadic_copy_constructible)
+      : __base_(__tuple_variadic_constructor_tag{}, __t...)
+  {}
+
+  template <class _Alloc, class _Constraints = __tuple_constraints<_Tp...>,
+      __enable_if_t<_Constraints::__implicit_variadic_copy_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(allocator_arg_t, const _Alloc& __a, const _Tp&... __t) noexcept(
+      _Constraints::__nothrow_variadic_copy_constructible)
+      : __base_(allocator_arg_t(), __a, __tuple_variadic_constructor_tag{}, __t...)
+  {}
+
+  template <class _Alloc, class _Constraints = __tuple_constraints<_Tp...>,
+      __enable_if_t<_Constraints::__explicit_variadic_copy_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY explicit tuple(allocator_arg_t, const _Alloc& __a, const _Tp&... __t) noexcept(
+      _Constraints::__nothrow_variadic_copy_constructible)
+      : __base_(allocator_arg_t(), __a, __tuple_variadic_constructor_tag{}, __t...)
+  {}
 
 #if defined(_LIBCUDACXX_NO_TUPLE_NOEXCEPT)
-    template <class ..._Vp>
-    using __base_noexcept_constructible = false_type;
+  template <class... _Vp>
+  using __base_noexcept_constructible = false_type;
 #else
-    template <class ..._Vp>
-    using __base_noexcept_constructible = is_nothrow_constructible<
-                _BaseT,
-                _Vp...
-            >;
+  template <class... _Vp>
+  using __base_noexcept_constructible = is_nothrow_constructible< _BaseT, _Vp... >;
 #endif // defined(_LIBCUDACXX_NO_TUPLE_NOEXCEPT)
 
-    template <class ..._Up>
-    struct _EnableVariadicConvertingConstructor {
-        static constexpr bool _PackIsTuple = _PackExpandsToThisTuple<_Up...>::value;
-        static constexpr bool __exact_match = sizeof...(_Up) == sizeof...(_Tp) && !_PackIsTuple;
-        static constexpr bool __less_arity  = sizeof...(_Up) < sizeof...(_Tp) && !_PackIsTuple;
-
-        static constexpr bool __implicit =
-            _CheckArgsConstructor<__exact_match>::template __enable_implicit<_Up...>::value ||
-            _CheckArgsConstructor<__less_arity && _EnableImplicitReducedArityExtension >::template __enable_implicit<_Up...>::value;
-        static constexpr bool __explicit =
-            _CheckArgsConstructor<__exact_match>::template __enable_explicit<_Up...>::value ||
-            _CheckArgsConstructor<__less_arity && !_EnableImplicitReducedArityExtension>::template __enable_implicit<_Up...>::value;
-
-        static constexpr bool __implicit_alloc =
-            _CheckArgsConstructor<__exact_match>::template __enable_implicit<_Up...>::value;
-        static constexpr bool __explicit_alloc =
-            _CheckArgsConstructor<__exact_match>::template __enable_explicit<_Up...>::value;
-    };
-    template <class ..._Up,
-              __enable_if_t<_EnableVariadicConvertingConstructor<_Up...>::__implicit, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        tuple(_Up&&... __u)
-            noexcept((
-                __base_noexcept_constructible<
-                    typename __make_tuple_indices<sizeof...(_Up)>::type,
-                    typename __make_tuple_types<tuple, sizeof...(_Up)>::type,
-                    typename __make_tuple_indices<sizeof...(_Tp), sizeof...(_Up)>::type,
-                    typename __make_tuple_types<tuple, sizeof...(_Tp), sizeof...(_Up)>::type,
-                    _Up...
-                >::value
-            ))
-            : __base_(typename __make_tuple_indices<sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Up)>::type(),
-                    typename __make_tuple_indices<sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    _CUDA_VSTD::forward<_Up>(__u)...) {}
-
-    template <class ..._Up,
-              __enable_if_t<_EnableVariadicConvertingConstructor<_Up...>::__explicit, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        explicit
-        tuple(_Up&&... __u)
-            noexcept((
-                __base_noexcept_constructible<
-                    typename __make_tuple_indices<sizeof...(_Up)>::type,
-                    typename __make_tuple_types<tuple, sizeof...(_Up)>::type,
-                    typename __make_tuple_indices<sizeof...(_Tp), sizeof...(_Up)>::type,
-                    typename __make_tuple_types<tuple, sizeof...(_Tp), sizeof...(_Up)>::type,
-                    _Up...
-                >::value
-            ))
-            : __base_(typename __make_tuple_indices<sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Up)>::type(),
-                    typename __make_tuple_indices<sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    _CUDA_VSTD::forward<_Up>(__u)...) {}
-
-    template <class _Alloc, class ..._Up,
-              __enable_if_t<_EnableVariadicConvertingConstructor<_Up...>::__implicit_alloc, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        tuple(allocator_arg_t, const _Alloc& __a, _Up&&... __u)
-            : __base_(allocator_arg_t(), __a,
-                    typename __make_tuple_indices<sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Up)>::type(),
-                    typename __make_tuple_indices<sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    _CUDA_VSTD::forward<_Up>(__u)...) {}
-
-    template <class _Alloc, class ..._Up,
-              __enable_if_t<_EnableVariadicConvertingConstructor<_Up...>::__explicit_alloc, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit
-        tuple(allocator_arg_t, const _Alloc& __a, _Up&&... __u)
-            : __base_(allocator_arg_t(), __a,
-                    typename __make_tuple_indices<sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Up)>::type(),
-                    typename __make_tuple_indices<sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    typename __make_tuple_types<tuple, sizeof...(_Tp), sizeof...(_Up)>::type(),
-                    _CUDA_VSTD::forward<_Up>(__u)...) {}
-
-
-    template <class _Tuple, bool _DisableIfLValue>
-    struct _EnableTupleLikeConstructor {
-        static constexpr bool __implicit =
-            _CheckTupleLikeConstructor<__tuple_like_with_size<_Tuple, sizeof...(_Tp)>::value
-                             && !_PackExpandsToThisTuple<_Tuple>::value
-                             && (!is_lvalue_reference<_Tuple>::value || !_DisableIfLValue)
-                         >::template __enable_implicit<_Tuple>::value;
-
-        static constexpr bool __explicit =
-            _CheckTupleLikeConstructor<__tuple_like_with_size<_Tuple, sizeof...(_Tp)>::value
-                             && !_PackExpandsToThisTuple<_Tuple>::value
-                             && (!is_lvalue_reference<_Tuple>::value || !_DisableIfLValue)
-                         >::template __enable_explicit<_Tuple>::value;
-
-        static constexpr bool __implicit_alloc =
-            _CheckTupleLikeConstructor<__tuple_like_with_size<_Tuple, sizeof...(_Tp)>::value>::template __enable_implicit<_Tuple>::value;
-
-        static constexpr bool __explicit_alloc =
-            _CheckTupleLikeConstructor<__tuple_like_with_size<_Tuple, sizeof...(_Tp)>::value>::template __enable_explicit<_Tuple>::value;
-    };
-
-    template <class _Tuple, __enable_if_t<_EnableTupleLikeConstructor<_Tuple, true>::__implicit, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        tuple(_Tuple&& __t) noexcept((is_nothrow_constructible<_BaseT, _Tuple>::value))
-            : __base_(_CUDA_VSTD::forward<_Tuple>(__t)) {}
-
-    template <class _Tuple, __enable_if_t<_EnableTupleLikeConstructor<const _Tuple&, false>::__implicit, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        tuple(const _Tuple& __t) noexcept((is_nothrow_constructible<_BaseT, const _Tuple&>::value))
-            : __base_(__t) {}
-
-    template <class _Tuple, __enable_if_t<_EnableTupleLikeConstructor<_Tuple, true>::__explicit, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        explicit
-        tuple(_Tuple&& __t) noexcept((is_nothrow_constructible<_BaseT, _Tuple>::value))
-            : __base_(_CUDA_VSTD::forward<_Tuple>(__t)) {}
-
-    template <class _Tuple, __enable_if_t<_EnableTupleLikeConstructor<const _Tuple&, false>::__explicit, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-        explicit
-        tuple(const _Tuple& __t) noexcept((is_nothrow_constructible<_BaseT, const _Tuple&>::value))
-            : __base_(__t) {}
-
-    template <class _Alloc, class _Tuple,
-              __enable_if_t<_EnableTupleLikeConstructor<_Tuple, true>::__implicit_alloc, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        tuple(allocator_arg_t, const _Alloc& __a, _Tuple&& __t)
-            : __base_(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tuple>(__t)) {}
-
-    template <class _Alloc, class _Tuple,
-              __enable_if_t<_EnableTupleLikeConstructor<_Tuple, true>::__explicit_alloc, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        explicit
-        tuple(allocator_arg_t, const _Alloc& __a, _Tuple&& __t)
-            : __base_(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tuple>(__t)) {}
-
-    using _CanCopyAssign = __all<is_copy_assignable<_Tp>::value...>;
-    using _CanMoveAssign = __all<is_move_assignable<_Tp>::value...>;
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    tuple& operator=(__conditional_t<_CanCopyAssign::value, tuple, __nat> const& __t)
-        noexcept((__all<is_nothrow_copy_assignable<_Tp>::value...>::value))
-    {
-        __base_.operator=(__t.__base_);
-        return *this;
-    }
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    tuple& operator=(__conditional_t<_CanMoveAssign::value, tuple, __nat>&& __t)
-        noexcept((__all<is_nothrow_move_assignable<_Tp>::value...>::value))
-    {
-        __base_.operator=(static_cast<_BaseT&&>(__t.__base_));
-        return *this;
-    }
-
-    template <class _Tuple,
-              __enable_if_t<__tuple_assignable<_Tuple, tuple>::value, bool> = false>
-        _LIBCUDACXX_INLINE_VISIBILITY
-        tuple&
-        operator=(_Tuple&& __t) noexcept((is_nothrow_assignable<_BaseT&, _Tuple>::value))
-        {
-            __base_.operator=(_CUDA_VSTD::forward<_Tuple>(__t));
-            return *this;
-        }
-
-    _LIBCUDACXX_INLINE_VISIBILITY
-    void swap(tuple& __t) noexcept(__all<__is_nothrow_swappable<_Tp>::value...>::value)
-        {__base_.swap(__t.__base_);}
+  template <class... _Up>
+  using __variadic_constraints = _If<!_PackExpandsToThisTuple<_Up...>::value && sizeof...(_Up) == sizeof...(_Tp),
+      typename __tuple_constraints<_Tp...>::template __variadic_constraints<_Up...>, __invalid_tuple_constraints>;
+
+  template <class... _Up, class _Constraints = __variadic_constraints<_Up...>,
+      __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
+  tuple(_Up&&... __u) noexcept(_Constraints::__nothrow_constructible)
+      : __base_(__tuple_variadic_constructor_tag{}, _CUDA_VSTD::forward<_Up>(__u)...)
+  {}
+
+  template <class... _Up, class _Constraints = __variadic_constraints<_Up...>,
+      __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit tuple(_Up&&... __u) noexcept(
+      _Constraints::__nothrow_constructible)
+      : __base_(__tuple_variadic_constructor_tag{}, _CUDA_VSTD::forward<_Up>(__u)...)
+  {}
+
+  template <class... _Up>
+  using __variadic_constraints_less_rank = _If<!_PackExpandsToThisTuple<_Up...>::value,
+      typename __tuple_constraints<_Tp...>::template __variadic_constraints_less_rank<_Up...>,
+      __invalid_tuple_constraints>;
+
+  template <class... _Up, class _Constraints = __variadic_constraints_less_rank<_Up...>,
+      __enable_if_t<sizeof...(_Up) < sizeof...(_Tp), int>        = 0,
+      __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit tuple(_Up&&... __u) noexcept(
+      __base_noexcept_constructible<__tuple_variadic_constructor_tag, _Up...>::value)
+      : __base_(__tuple_variadic_constructor_tag{}, _CUDA_VSTD::forward<_Up>(__u)...)
+  {}
+
+  template <class _Alloc, class... _Up, class _Constraints = __variadic_constraints<_Up...>,
+      __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(allocator_arg_t, const _Alloc& __a, _Up&&... __u) noexcept(_Constraints::__nothrow_constructible)
+      : __base_(allocator_arg_t(), __a, __tuple_variadic_constructor_tag{}, _CUDA_VSTD::forward<_Up>(__u)...)
+  {}
+
+  template <class _Alloc, class... _Up, class _Constraints = __variadic_constraints<_Up...>,
+      __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  explicit tuple(allocator_arg_t, const _Alloc& __a, _Up&&... __u) noexcept(
+      _Constraints::__nothrow_constructible)
+      : __base_(allocator_arg_t(), __a, __tuple_variadic_constructor_tag{}, _CUDA_VSTD::forward<_Up>(__u)...)
+  {}
+
+  template <class _Tuple>
+  using __tuple_like_constraints = _If<__tuple_like_with_size<_Tuple, sizeof...(_Tp)>::value,
+      typename __tuple_constraints< _Tp...>::template __tuple_like_constraints<_Tuple>, __invalid_tuple_constraints>;
+
+  template < class _Tuple, class _Constraints = __tuple_like_constraints<_Tuple>,
+      __enable_if_t<!_PackExpandsToThisTuple<_Tuple>::value, int>         = 0,
+      __enable_if_t<!_LIBCUDACXX_TRAIT(is_lvalue_reference, _Tuple), int> = 0,
+      __enable_if_t<_Constraints::__implicit_constructible, int>          = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
+  tuple(_Tuple&& __t) noexcept(_LIBCUDACXX_TRAIT(is_nothrow_constructible, _BaseT, _Tuple))
+      : __base_(_CUDA_VSTD::forward<_Tuple>(__t))
+  {}
+
+  template <class _Tuple, class _Constraints = __tuple_like_constraints<const _Tuple&>,
+      __enable_if_t<!_PackExpandsToThisTuple<_Tuple>::value, int> = 0,
+      __enable_if_t<_Constraints::__implicit_constructible, int>  = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
+  tuple(const _Tuple& __t) noexcept(_LIBCUDACXX_TRAIT(is_nothrow_constructible, _BaseT, const _Tuple&))
+      : __base_(__t)
+  {}
+
+  template < class _Tuple, class _Constraints = __tuple_like_constraints<_Tuple>,
+      __enable_if_t<!_PackExpandsToThisTuple<_Tuple>::value, int>         = 0,
+      __enable_if_t<!_LIBCUDACXX_TRAIT(is_lvalue_reference, _Tuple), int> = 0,
+      __enable_if_t<_Constraints::__explicit_constructible, int>          = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit tuple(_Tuple&& __t) noexcept(
+      _LIBCUDACXX_TRAIT(is_nothrow_constructible, _BaseT, _Tuple))
+      : __base_(_CUDA_VSTD::forward<_Tuple>(__t))
+  {}
+
+  template <class _Tuple, class _Constraints = __tuple_like_constraints<const _Tuple&>,
+      __enable_if_t<!_PackExpandsToThisTuple<_Tuple>::value, int> = 0,
+      __enable_if_t<_Constraints::__explicit_constructible, int>  = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 explicit tuple(const _Tuple& __t) noexcept(
+      _LIBCUDACXX_TRAIT(is_nothrow_constructible, _BaseT, const _Tuple&))
+      : __base_(__t)
+  {}
+
+  template <class _Alloc, class _Tuple, class _Constraints = __tuple_like_constraints<_Tuple>,
+      __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(allocator_arg_t, const _Alloc& __a, _Tuple&& __t)
+      : __base_(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tuple>(__t))
+  {}
+
+  template <class _Alloc, class _Tuple, class _Constraints = __tuple_like_constraints<_Tuple>,
+      __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  explicit tuple(allocator_arg_t, const _Alloc& __a, _Tuple&& __t)
+      : __base_(allocator_arg_t(), __a, _CUDA_VSTD::forward<_Tuple>(__t))
+  {}
+
+  using _CanCopyAssign = __all<_LIBCUDACXX_TRAIT(is_copy_assignable, _Tp)...>;
+  using _CanMoveAssign = __all<_LIBCUDACXX_TRAIT(is_move_assignable, _Tp)...>;
+
+  _LIBCUDACXX_INLINE_VISIBILITY tuple&
+  operator=(__conditional_t<_CanCopyAssign::value, tuple, __nat> const& __t) noexcept(
+      (__all<_LIBCUDACXX_TRAIT(is_nothrow_copy_assignable, _Tp)...>::value))
+  {
+    __base_.operator=(__t.__base_);
+    return *this;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY tuple&
+  operator=(__conditional_t<_CanMoveAssign::value, tuple, __nat>&& __t) noexcept(
+      (__all<_LIBCUDACXX_TRAIT(is_nothrow_move_assignable, _Tp)...>::value))
+  {
+    __base_.operator=(static_cast<_BaseT&&>(__t.__base_));
+    return *this;
+  }
+
+  template < class _Tuple, __enable_if_t<__tuple_assignable<_Tuple, tuple>::value, bool> = false>
+  _LIBCUDACXX_INLINE_VISIBILITY tuple&
+  operator=(_Tuple&& __t) noexcept(_LIBCUDACXX_TRAIT(is_nothrow_assignable, _BaseT&, _Tuple))
+  {
+    __base_.operator=(_CUDA_VSTD::forward<_Tuple>(__t));
+    return *this;
+  }
+
+  _LIBCUDACXX_INLINE_VISIBILITY void
+  swap(tuple& __t) noexcept(__all<__is_nothrow_swappable<_Tp>::value...>::value)
+  {
+    __base_.swap(__t.__base_);
+  }
 };
 
 template <>
 class _LIBCUDACXX_TEMPLATE_VIS tuple<>
 {
 public:
-    constexpr tuple() noexcept = default;
-    template <class _Alloc>
-    _LIBCUDACXX_INLINE_VISIBILITY
-        tuple(allocator_arg_t, const _Alloc&) noexcept {}
-    template <class _Alloc>
-    _LIBCUDACXX_INLINE_VISIBILITY
-        tuple(allocator_arg_t, const _Alloc&, const tuple&) noexcept {}
-    template <class _Up>
-    _LIBCUDACXX_INLINE_VISIBILITY
-        tuple(array<_Up, 0>) noexcept {}
-    template <class _Alloc, class _Up>
-    _LIBCUDACXX_INLINE_VISIBILITY
-        tuple(allocator_arg_t, const _Alloc&, array<_Up, 0>) noexcept {}
-    _LIBCUDACXX_INLINE_VISIBILITY
-    void swap(tuple&) noexcept {}
+  constexpr tuple() noexcept = default;
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(allocator_arg_t, const _Alloc&) noexcept
+  {}
+  template <class _Alloc>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(allocator_arg_t, const _Alloc&, const tuple&) noexcept
+  {}
+  template <class _Up>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(array<_Up, 0>) noexcept
+  {}
+  template <class _Alloc, class _Up>
+  _LIBCUDACXX_INLINE_VISIBILITY
+  tuple(allocator_arg_t, const _Alloc&, array<_Up, 0>) noexcept
+  {}
+  _LIBCUDACXX_INLINE_VISIBILITY void
+  swap(tuple&) noexcept
+  {}
 };
 
 #ifndef _LIBCUDACXX_HAS_NO_DEDUCTION_GUIDES
-template <class ..._Tp>
+template <class... _Tp>
 tuple(_Tp...) -> tuple<_Tp...>;
 template <class _Tp1, class _Tp2>
 tuple(pair<_Tp1, _Tp2>) -> tuple<_Tp1, _Tp2>;
-template <class _Alloc, class ..._Tp>
+template <class _Alloc, class... _Tp>
 tuple(allocator_arg_t, _Alloc, _Tp...) -> tuple<_Tp...>;
 template <class _Alloc, class _Tp1, class _Tp2>
 tuple(allocator_arg_t, _Alloc, pair<_Tp1, _Tp2>) -> tuple<_Tp1, _Tp2>;
-template <class _Alloc, class ..._Tp>
+template <class _Alloc, class... _Tp>
 tuple(allocator_arg_t, _Alloc, tuple<_Tp...>) -> tuple<_Tp...>;
 #endif
 
-template <class ..._Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-__enable_if_t<_And<__is_swappable<_Tp>...>::value, void>
-swap(tuple<_Tp...>& __t, tuple<_Tp...>& __u) noexcept(__all<__is_nothrow_swappable<_Tp>::value...>::value) {
-    __t.swap(__u);
+template <class... _Tp>
+inline _LIBCUDACXX_INLINE_VISIBILITY __enable_if_t<_And<__is_swappable<_Tp>...>::value, void>
+swap(tuple<_Tp...>& __t, tuple<_Tp...>& __u) noexcept(__all<__is_nothrow_swappable<_Tp>::value...>::value)
+{
+  __t.swap(__u);
 }
 
 // get
-template <size_t _Ip, class ..._Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, tuple<_Tp...> >::type&
+template <size_t _Ip, class... _Tp>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, tuple<_Tp...>>&
 get(tuple<_Tp...>& __t) noexcept
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename tuple_element<_Ip, tuple<_Tp...> >::type type;
-    return static_cast<__tuple_leaf<_Ip, type>&>(__t.__base_).get();
+  typedef _LIBCUDACXX_NODEBUG_TYPE __tuple_element_t<_Ip, tuple<_Tp...>> type;
+  return static_cast<__tuple_leaf<_Ip, type>&>(__t.__base_).get();
 }
 
-template <size_t _Ip, class ..._Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, tuple<_Tp...> >::type&
+template <size_t _Ip, class... _Tp>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, tuple<_Tp...>>&
 get(const tuple<_Tp...>& __t) noexcept
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename tuple_element<_Ip, tuple<_Tp...> >::type type;
-    return static_cast<const __tuple_leaf<_Ip, type>&>(__t.__base_).get();
+  typedef _LIBCUDACXX_NODEBUG_TYPE __tuple_element_t<_Ip, tuple<_Tp...>> type;
+  return static_cast<const __tuple_leaf<_Ip, type>&>(__t.__base_).get();
 }
 
-template <size_t _Ip, class ..._Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename tuple_element<_Ip, tuple<_Tp...> >::type&&
+template <size_t _Ip, class... _Tp>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 __tuple_element_t<_Ip, tuple<_Tp...>>&&
 get(tuple<_Tp...>&& __t) noexcept
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename tuple_element<_Ip, tuple<_Tp...> >::type type;
-    return static_cast<type&&>(
-             static_cast<__tuple_leaf<_Ip, type>&&>(__t.__base_).get());
+  typedef _LIBCUDACXX_NODEBUG_TYPE __tuple_element_t<_Ip, tuple<_Tp...>> type;
+  return static_cast<type&&>(static_cast<__tuple_leaf<_Ip, type>&&>(__t.__base_).get());
 }
 
-template <size_t _Ip, class ..._Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-const typename tuple_element<_Ip, tuple<_Tp...> >::type&&
+template <size_t _Ip, class... _Tp>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __tuple_element_t<_Ip, tuple<_Tp...>>&&
 get(const tuple<_Tp...>&& __t) noexcept
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename tuple_element<_Ip, tuple<_Tp...> >::type type;
-    return static_cast<const type&&>(
-             static_cast<const __tuple_leaf<_Ip, type>&&>(__t.__base_).get());
+  typedef _LIBCUDACXX_NODEBUG_TYPE __tuple_element_t<_Ip, tuple<_Tp...>> type;
+  return static_cast<const type&&>(static_cast<const __tuple_leaf<_Ip, type>&&>(__t.__base_).get());
 }
 
 #if _LIBCUDACXX_STD_VER > 11
@@ -1093,412 +954,373 @@ namespace __find_detail {
 static constexpr size_t __not_found = ~size_t(0);
 static constexpr size_t __ambiguous = __not_found - 1;
 
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr size_t __find_idx_return(size_t __curr_i, size_t __res, bool __matches) {
-    return !__matches ? __res :
-        (__res == __not_found ? __curr_i : __ambiguous);
+inline _LIBCUDACXX_INLINE_VISIBILITY constexpr size_t
+__find_idx_return(size_t __curr_i, size_t __res, bool __matches)
+{
+  return !__matches ? __res : (__res == __not_found ? __curr_i : __ambiguous);
 }
 
 template <size_t _Nx>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr size_t __find_idx(size_t __i, const bool (&__matches)[_Nx]) {
-  return __i == _Nx ? __not_found :
-      __find_idx_return(__i, __find_idx(__i + 1, __matches), __matches[__i]);
+inline _LIBCUDACXX_INLINE_VISIBILITY constexpr size_t
+__find_idx(size_t __i, const bool (&__matches)[_Nx])
+{
+  return __i == _Nx ? __not_found : __find_idx_return(__i, __find_idx(__i + 1, __matches), __matches[__i]);
 }
 
-template <class _T1, class ..._Args>
-struct __find_exactly_one_checked {
-    static constexpr bool __matches[sizeof...(_Args)] = {is_same<_T1, _Args>::value...};
-    static constexpr size_t value = __find_detail::__find_idx(0, __matches);
-    static_assert(value != __not_found, "type not found in type list" );
-    static_assert(value != __ambiguous, "type occurs more than once in type list");
+template <class _T1, class... _Args>
+struct __find_exactly_one_checked
+{
+  static constexpr bool __matches[sizeof...(_Args)] = {is_same<_T1, _Args>::value...};
+  static constexpr size_t value                     = __find_detail::__find_idx(0, __matches);
+  static_assert(value != __not_found, "type not found in type list");
+  static_assert(value != __ambiguous, "type occurs more than once in type list");
 };
 
 template <class _T1>
-struct __find_exactly_one_checked<_T1> {
-    static_assert(!is_same<_T1, _T1>::value, "type not in empty type list");
+struct __find_exactly_one_checked<_T1>
+{
+  static_assert(!is_same<_T1, _T1>::value, "type not in empty type list");
 };
 
-} // namespace __find_detail;
+} // namespace __find_detail
 
 template <typename _T1, typename... _Args>
-struct __find_exactly_one_t
-    : public __find_detail::__find_exactly_one_checked<_T1, _Args...> {
-};
+struct __find_exactly_one_t : public __find_detail::__find_exactly_one_checked<_T1, _Args...>
+{};
 
 template <class _T1, class... _Args>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr _T1& get(tuple<_Args...>& __tup) noexcept
+inline _LIBCUDACXX_INLINE_VISIBILITY constexpr _T1&
+get(tuple<_Args...>& __tup) noexcept
 {
-    return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(__tup);
+  return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(__tup);
 }
 
 template <class _T1, class... _Args>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr _T1 const& get(tuple<_Args...> const& __tup) noexcept
+inline _LIBCUDACXX_INLINE_VISIBILITY constexpr _T1 const&
+get(tuple<_Args...> const& __tup) noexcept
 {
-    return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(__tup);
+  return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(__tup);
 }
 
 template <class _T1, class... _Args>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr _T1&& get(tuple<_Args...>&& __tup) noexcept
+inline _LIBCUDACXX_INLINE_VISIBILITY constexpr _T1&&
+get(tuple<_Args...>&& __tup) noexcept
 {
-    return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(_CUDA_VSTD::move(__tup));
+  return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(_CUDA_VSTD::move(__tup));
 }
 
 template <class _T1, class... _Args>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr _T1 const&& get(tuple<_Args...> const&& __tup) noexcept
+inline _LIBCUDACXX_INLINE_VISIBILITY constexpr _T1 const&&
+get(tuple<_Args...> const&& __tup) noexcept
 {
-    return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(_CUDA_VSTD::move(__tup));
+  return _CUDA_VSTD::get<__find_exactly_one_t<_T1, _Args...>::value>(_CUDA_VSTD::move(__tup));
 }
 
 #endif
 
 // tie
 
-template <class ..._Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-tuple<_Tp&...>
+template <class... _Tp>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 tuple<_Tp&...>
 tie(_Tp&... __t) noexcept
 {
-    return tuple<_Tp&...>(__t...);
+  return tuple<_Tp&...>(__t...);
 }
 
 template <class _Up>
 struct __ignore_t
 {
-    template <class _Tp>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    const __ignore_t& operator=(_Tp&&) const {return *this;}
+  template <class _Tp>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 const __ignore_t&
+  operator=(_Tp&&) const
+  {
+    return *this;
+  }
 };
 
 namespace {
 #ifdef __CUDA_ARCH__
-    static
+static
 #endif
-    _LIBCUDACXX_INLINE_VAR constexpr __ignore_t<unsigned char> ignore = __ignore_t<unsigned char>();
-}
+    _LIBCUDACXX_INLINE_VAR constexpr __ignore_t<unsigned char>
+        ignore = __ignore_t<unsigned char>();
+} // namespace
 
 template <class... _Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-tuple<typename __unwrap_ref_decay<_Tp>::type...>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 tuple<typename __unwrap_ref_decay<_Tp>::type...>
 make_tuple(_Tp&&... __t)
 {
-    return tuple<typename __unwrap_ref_decay<_Tp>::type...>(_CUDA_VSTD::forward<_Tp>(__t)...);
+  return tuple<typename __unwrap_ref_decay<_Tp>::type...>(_CUDA_VSTD::forward<_Tp>(__t)...);
 }
 
 template <class... _Tp>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-tuple<_Tp&&...>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 tuple<_Tp&&...>
 forward_as_tuple(_Tp&&... __t) noexcept
 {
-    return tuple<_Tp&&...>(_CUDA_VSTD::forward<_Tp>(__t)...);
+  return tuple<_Tp&&...>(_CUDA_VSTD::forward<_Tp>(__t)...);
 }
 
 template <size_t _Ip>
 struct __tuple_equal
 {
-    template <class _Tp, class _Up>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    bool operator()(const _Tp& __x, const _Up& __y)
-    {
-        return __tuple_equal<_Ip - 1>()(__x, __y) && _CUDA_VSTD::get<_Ip-1>(__x) == _CUDA_VSTD::get<_Ip-1>(__y);
-    }
+  template <class _Tp, class _Up>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
+  operator()(const _Tp& __x, const _Up& __y)
+  {
+    return __tuple_equal<_Ip - 1>()(__x, __y) && _CUDA_VSTD::get<_Ip - 1>(__x) == _CUDA_VSTD::get<_Ip - 1>(__y);
+  }
 };
 
 template <>
 struct __tuple_equal<0>
 {
-    template <class _Tp, class _Up>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    bool operator()(const _Tp&, const _Up&)
-    {
-        return true;
-    }
+  template <class _Tp, class _Up>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
+  operator()(const _Tp&, const _Up&)
+  {
+    return true;
+  }
 };
 
-template <class ..._Tp, class ..._Up>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-bool
+template <class... _Tp, class... _Up>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 operator==(const tuple<_Tp...>& __x, const tuple<_Up...>& __y)
 {
-    static_assert (sizeof...(_Tp) == sizeof...(_Up), "Can't compare tuples of different sizes");
-    return __tuple_equal<sizeof...(_Tp)>()(__x, __y);
+  static_assert(sizeof...(_Tp) == sizeof...(_Up), "Can't compare tuples of different sizes");
+  return __tuple_equal<sizeof...(_Tp)>()(__x, __y);
 }
 
-template <class ..._Tp, class ..._Up>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-bool
+template <class... _Tp, class... _Up>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 operator!=(const tuple<_Tp...>& __x, const tuple<_Up...>& __y)
 {
-    return !(__x == __y);
+  return !(__x == __y);
 }
 
 template <size_t _Ip>
 struct __tuple_less
 {
-    template <class _Tp, class _Up>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    bool operator()(const _Tp& __x, const _Up& __y)
-    {
-        const size_t __idx = tuple_size<_Tp>::value - _Ip;
-        if (_CUDA_VSTD::get<__idx>(__x) < _CUDA_VSTD::get<__idx>(__y))
-            return true;
-        if (_CUDA_VSTD::get<__idx>(__y) < _CUDA_VSTD::get<__idx>(__x))
-            return false;
-        return __tuple_less<_Ip-1>()(__x, __y);
+  template <class _Tp, class _Up>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
+  operator()(const _Tp& __x, const _Up& __y)
+  {
+    const size_t __idx = tuple_size<_Tp>::value - _Ip;
+    if (_CUDA_VSTD::get<__idx>(__x) < _CUDA_VSTD::get<__idx>(__y)) {
+      return true;
+    }
+    if (_CUDA_VSTD::get<__idx>(__y) < _CUDA_VSTD::get<__idx>(__x)) {
+      return false;
     }
+    return __tuple_less<_Ip - 1>()(__x, __y);
+  }
 };
 
 template <>
 struct __tuple_less<0>
 {
-    template <class _Tp, class _Up>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    bool operator()(const _Tp&, const _Up&)
-    {
-        return false;
-    }
+  template <class _Tp, class _Up>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
+  operator()(const _Tp&, const _Up&)
+  {
+    return false;
+  }
 };
 
-template <class ..._Tp, class ..._Up>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-bool
+template <class... _Tp, class... _Up>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 operator<(const tuple<_Tp...>& __x, const tuple<_Up...>& __y)
 {
-    static_assert (sizeof...(_Tp) == sizeof...(_Up), "Can't compare tuples of different sizes");
-    return __tuple_less<sizeof...(_Tp)>()(__x, __y);
+  static_assert(sizeof...(_Tp) == sizeof...(_Up), "Can't compare tuples of different sizes");
+  return __tuple_less<sizeof...(_Tp)>()(__x, __y);
 }
 
-template <class ..._Tp, class ..._Up>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-bool
+template <class... _Tp, class... _Up>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 operator>(const tuple<_Tp...>& __x, const tuple<_Up...>& __y)
 {
-    return __y < __x;
+  return __y < __x;
 }
 
-template <class ..._Tp, class ..._Up>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-bool
+template <class... _Tp, class... _Up>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 operator>=(const tuple<_Tp...>& __x, const tuple<_Up...>& __y)
 {
-    return !(__x < __y);
+  return !(__x < __y);
 }
 
-template <class ..._Tp, class ..._Up>
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-bool
+template <class... _Tp, class... _Up>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 operator<=(const tuple<_Tp...>& __x, const tuple<_Up...>& __y)
 {
-    return !(__y < __x);
+  return !(__y < __x);
 }
 
 // tuple_cat
 
-template <class _Tp, class _Up> struct __tuple_cat_type;
+template <class _Tp, class _Up>
+struct __tuple_cat_type;
 
-template <class ..._Ttypes, class ..._Utypes>
-struct __tuple_cat_type<tuple<_Ttypes...>, __tuple_types<_Utypes...> >
+template <class... _Ttypes, class... _Utypes>
+struct __tuple_cat_type<tuple<_Ttypes...>, __tuple_types<_Utypes...>>
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE tuple<_Ttypes..., _Utypes...> type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE tuple<_Ttypes..., _Utypes...> type;
 };
 
-template <class _ResultTuple, bool _Is_Tuple0TupleLike, class ..._Tuples>
+template <class _ResultTuple, bool _Is_Tuple0TupleLike, class... _Tuples>
 struct __tuple_cat_return_1
-{
-};
+{};
 
-template <class ..._Types, class _Tuple0>
+template <class... _Types, class _Tuple0>
 struct __tuple_cat_return_1<tuple<_Types...>, true, _Tuple0>
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename __tuple_cat_type<tuple<_Types...>,
-            typename __make_tuple_types<__remove_cvref_t<_Tuple0>>::type>::type
-                                                                           type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE
+      typename __tuple_cat_type< tuple<_Types...>, __make_tuple_types_t<__remove_cvref_t<_Tuple0>>>::type type;
 };
 
-template <class ..._Types, class _Tuple0, class _Tuple1, class ..._Tuples>
+template <class... _Types, class _Tuple0, class _Tuple1, class... _Tuples>
 struct __tuple_cat_return_1<tuple<_Types...>, true, _Tuple0, _Tuple1, _Tuples...>
     : public __tuple_cat_return_1<
-                 typename __tuple_cat_type<
-                     tuple<_Types...>,
-                     typename __make_tuple_types<__remove_cvref_t<_Tuple0>>::type
-                 >::type,
-                 __tuple_like<typename remove_reference<_Tuple1>::type>::value,
-                 _Tuple1, _Tuples...>
-{
-};
+          typename __tuple_cat_type< tuple<_Types...>, __make_tuple_types_t<__remove_cvref_t<_Tuple0>>>::type,
+          __tuple_like<__libcpp_remove_reference_t<_Tuple1>>::value, _Tuple1, _Tuples...>
+{};
 
-template <class ..._Tuples> struct __tuple_cat_return;
+template <class... _Tuples>
+struct __tuple_cat_return;
 
-template <class _Tuple0, class ..._Tuples>
+template <class _Tuple0, class... _Tuples>
 struct __tuple_cat_return<_Tuple0, _Tuples...>
-    : public __tuple_cat_return_1<tuple<>,
-         __tuple_like<typename remove_reference<_Tuple0>::type>::value, _Tuple0,
-                                                                     _Tuples...>
-{
-};
+    : public __tuple_cat_return_1< tuple<>, __tuple_like<__libcpp_remove_reference_t<_Tuple0>>::value, _Tuple0,
+          _Tuples...>
+{};
 
 template <>
 struct __tuple_cat_return<>
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE tuple<> type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE tuple<> type;
 };
 
-inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-tuple<>
+inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 tuple<>
 tuple_cat()
 {
-    return tuple<>();
+  return tuple<>();
 }
 
-template <class _Rp, class _Indices, class _Tuple0, class ..._Tuples>
+template <class _Rp, class _Indices, class _Tuple0, class... _Tuples>
 struct __tuple_cat_return_ref_imp;
 
-template <class ..._Types, size_t ..._I0, class _Tuple0>
+template <class... _Types, size_t... _I0, class _Tuple0>
 struct __tuple_cat_return_ref_imp<tuple<_Types...>, __tuple_indices<_I0...>, _Tuple0>
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename remove_reference<_Tuple0>::type _T0;
-    typedef tuple<_Types..., typename __apply_cv<_Tuple0,
-                          typename tuple_element<_I0, _T0>::type>::type&&...> type;
+  typedef _LIBCUDACXX_NODEBUG_TYPE __libcpp_remove_reference_t<_Tuple0> _T0;
+  typedef tuple< _Types..., typename __apply_cv<_Tuple0, __tuple_element_t<_I0, _T0>>::type&&...> type;
 };
 
-template <class ..._Types, size_t ..._I0, class _Tuple0, class _Tuple1, class ..._Tuples>
-struct __tuple_cat_return_ref_imp<tuple<_Types...>, __tuple_indices<_I0...>,
-                                  _Tuple0, _Tuple1, _Tuples...>
+template <class... _Types, size_t... _I0, class _Tuple0, class _Tuple1, class... _Tuples>
+struct __tuple_cat_return_ref_imp<tuple<_Types...>, __tuple_indices<_I0...>, _Tuple0, _Tuple1, _Tuples...>
     : public __tuple_cat_return_ref_imp<
-         tuple<_Types..., typename __apply_cv<_Tuple0,
-               typename tuple_element<_I0,
-                  typename remove_reference<_Tuple0>::type>::type>::type&&...>,
-         typename __make_tuple_indices<tuple_size<typename
-                                 remove_reference<_Tuple1>::type>::value>::type,
-         _Tuple1, _Tuples...>
-{
-};
+          tuple<_Types...,
+              typename __apply_cv< _Tuple0, __tuple_element_t<_I0, __libcpp_remove_reference_t< _Tuple0>>>::type&&...>,
+          __make_tuple_indices_t< tuple_size<__libcpp_remove_reference_t<_Tuple1>>::value>, _Tuple1, _Tuples...>
+{};
 
-template <class _Tuple0, class ..._Tuples>
+template <class _Tuple0, class... _Tuples>
 struct __tuple_cat_return_ref
-    : public __tuple_cat_return_ref_imp<tuple<>,
-               typename __make_tuple_indices<
-                        tuple_size<typename remove_reference<_Tuple0>::type>::value
-               >::type, _Tuple0, _Tuples...>
-{
-};
+    : public __tuple_cat_return_ref_imp< tuple<>,
+          __make_tuple_indices_t< tuple_size<__libcpp_remove_reference_t<_Tuple0>>::value>, _Tuple0, _Tuples...>
+{};
 
 template <class _Types, class _I0, class _J0>
 struct __tuple_cat;
 
-template <class ..._Types, size_t ..._I0, size_t ..._J0>
-struct __tuple_cat<tuple<_Types...>, __tuple_indices<_I0...>, __tuple_indices<_J0...> >
+template <class... _Types, size_t... _I0, size_t... _J0>
+struct __tuple_cat<tuple<_Types...>, __tuple_indices<_I0...>, __tuple_indices<_J0...>>
 {
-    template <class _Tuple0>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    typename __tuple_cat_return_ref<tuple<_Types...>&&, _Tuple0&&>::type
-    operator()(tuple<_Types...> __t, _Tuple0&& __t0)
-    {
-        (void)__t;
-        return _CUDA_VSTD::forward_as_tuple(_CUDA_VSTD::forward<_Types>(_CUDA_VSTD::get<_I0>(__t))...,
-                                            _CUDA_VSTD::get<_J0>(_CUDA_VSTD::forward<_Tuple0>(__t0))...);
-    }
-
-    template <class _Tuple0, class _Tuple1, class ..._Tuples>
-    _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-    typename __tuple_cat_return_ref<tuple<_Types...>&&, _Tuple0&&, _Tuple1&&, _Tuples&&...>::type
-    operator()(tuple<_Types...> __t, _Tuple0&& __t0, _Tuple1&& __t1, _Tuples&& ...__tpls)
-    {
-        (void)__t;
-        typedef _LIBCUDACXX_NODEBUG_TYPE typename remove_reference<_Tuple0>::type _T0;
-        typedef _LIBCUDACXX_NODEBUG_TYPE typename remove_reference<_Tuple1>::type _T1;
-        return __tuple_cat<
-           tuple<_Types..., typename __apply_cv<_Tuple0, typename tuple_element<_J0, _T0>::type>::type&&...>,
-           typename __make_tuple_indices<sizeof ...(_Types) + tuple_size<_T0>::value>::type,
-           typename __make_tuple_indices<tuple_size<_T1>::value>::type>()
-                           (_CUDA_VSTD::forward_as_tuple(
-                              _CUDA_VSTD::forward<_Types>(_CUDA_VSTD::get<_I0>(__t))...,
-                              _CUDA_VSTD::get<_J0>(_CUDA_VSTD::forward<_Tuple0>(__t0))...
-                            ),
-                            _CUDA_VSTD::forward<_Tuple1>(__t1),
-                            _CUDA_VSTD::forward<_Tuples>(__tpls)...);
-    }
+  template <class _Tuple0>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
+      typename __tuple_cat_return_ref<tuple<_Types...>&&, _Tuple0&&>::type
+      operator()(tuple<_Types...> __t, _Tuple0&& __t0)
+  {
+    (void) __t;
+    return _CUDA_VSTD::forward_as_tuple(_CUDA_VSTD::forward<_Types>(_CUDA_VSTD::get<_I0>(__t))...,
+        _CUDA_VSTD::get<_J0>(_CUDA_VSTD::forward<_Tuple0>(__t0))...);
+  }
+
+  template <class _Tuple0, class _Tuple1, class... _Tuples>
+  _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
+      typename __tuple_cat_return_ref<tuple<_Types...>&&, _Tuple0&&, _Tuple1&&, _Tuples&&...>::type
+      operator()(tuple<_Types...> __t, _Tuple0&& __t0, _Tuple1&& __t1, _Tuples&&... __tpls)
+  {
+    (void) __t;
+    typedef _LIBCUDACXX_NODEBUG_TYPE __libcpp_remove_reference_t<_Tuple0> _T0;
+    typedef _LIBCUDACXX_NODEBUG_TYPE __libcpp_remove_reference_t<_Tuple1> _T1;
+    return __tuple_cat< tuple<_Types..., typename __apply_cv< _Tuple0, __tuple_element_t<_J0, _T0>>::type&&...>,
+        __make_tuple_indices_t<sizeof...(_Types) + tuple_size<_T0>::value>,
+        __make_tuple_indices_t<tuple_size<_T1>::value>>()(
+        _CUDA_VSTD::forward_as_tuple(_CUDA_VSTD::forward<_Types>(_CUDA_VSTD::get<_I0>(__t))...,
+            _CUDA_VSTD::get<_J0>(_CUDA_VSTD::forward<_Tuple0>(__t0))...),
+        _CUDA_VSTD::forward<_Tuple1>(__t1),
+        _CUDA_VSTD::forward<_Tuples>(__tpls)...);
+  }
 };
 
 template <class _Tuple0, class... _Tuples>
 inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
-typename __tuple_cat_return<_Tuple0, _Tuples...>::type
-tuple_cat(_Tuple0&& __t0, _Tuples&&... __tpls)
+    typename __tuple_cat_return<_Tuple0, _Tuples...>::type
+    tuple_cat(_Tuple0&& __t0, _Tuples&&... __tpls)
 {
-    typedef _LIBCUDACXX_NODEBUG_TYPE typename remove_reference<_Tuple0>::type _T0;
-    return __tuple_cat<tuple<>, __tuple_indices<>,
-                  typename __make_tuple_indices<tuple_size<_T0>::value>::type>()
-                  (tuple<>(), _CUDA_VSTD::forward<_Tuple0>(__t0),
-                                            _CUDA_VSTD::forward<_Tuples>(__tpls)...);
+  typedef _LIBCUDACXX_NODEBUG_TYPE __libcpp_remove_reference_t<_Tuple0> _T0;
+  return __tuple_cat<tuple<>, __tuple_indices<>, __make_tuple_indices_t<tuple_size<_T0>::value>>()(
+      tuple<>(), _CUDA_VSTD::forward<_Tuple0>(__t0), _CUDA_VSTD::forward<_Tuples>(__tpls)...);
 }
 
-template <class ..._Tp, class _Alloc>
-struct _LIBCUDACXX_TEMPLATE_VIS uses_allocator<tuple<_Tp...>, _Alloc>
-    : true_type {};
+template <class... _Tp, class _Alloc>
+struct _LIBCUDACXX_TEMPLATE_VIS uses_allocator<tuple<_Tp...>, _Alloc> : true_type
+{};
 
 template <class _T1, class _T2>
-template <class... _Args1, class... _Args2, size_t ..._I1, size_t ..._I2>
+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))...)
-{
-}
+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))...)
+{}
 
 #if _LIBCUDACXX_STD_VER > 14
-template <class _Tp>
-_LIBCUDACXX_INLINE_VAR constexpr size_t tuple_size_v = tuple_size<_Tp>::value;
-
-#define _LIBCUDACXX_NOEXCEPT_RETURN(...) noexcept(noexcept(__VA_ARGS__)) { return __VA_ARGS__; }
-
-template <class _Fn, class _Tuple, size_t ..._Id>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr decltype(auto) __apply_tuple_impl(_Fn && __f, _Tuple && __t,
-                                            __tuple_indices<_Id...>)
-_LIBCUDACXX_NOEXCEPT_RETURN(
-    _CUDA_VSTD::__invoke(
-        _CUDA_VSTD::forward<_Fn>(__f),
-        _CUDA_VSTD::get<_Id>(_CUDA_VSTD::forward<_Tuple>(__t))...)
-)
-
-template <class _Fn, class _Tuple>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr decltype(auto) apply(_Fn && __f, _Tuple && __t)
-_LIBCUDACXX_NOEXCEPT_RETURN(
-    _CUDA_VSTD::__apply_tuple_impl(
-        _CUDA_VSTD::forward<_Fn>(__f), _CUDA_VSTD::forward<_Tuple>(__t),
-        typename __make_tuple_indices<tuple_size_v<remove_reference_t<_Tuple>>>::type{})
-)
-
-template <class _Tp, class _Tuple, size_t... _Idx>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr _Tp __make_from_tuple_impl(_Tuple&& __t, __tuple_indices<_Idx...>)
-_LIBCUDACXX_NOEXCEPT_RETURN(
-    _Tp(_CUDA_VSTD::get<_Idx>(_CUDA_VSTD::forward<_Tuple>(__t))...)
-)
-
-template <class _Tp, class _Tuple>
-inline _LIBCUDACXX_INLINE_VISIBILITY
-constexpr _Tp make_from_tuple(_Tuple&& __t)
-_LIBCUDACXX_NOEXCEPT_RETURN(
-    _CUDA_VSTD::__make_from_tuple_impl<_Tp>(_CUDA_VSTD::forward<_Tuple>(__t),
-        typename __make_tuple_indices<tuple_size_v<remove_reference_t<_Tuple>>>::type{})
-)
-
-#undef _LIBCUDACXX_NOEXCEPT_RETURN
+#  define _LIBCUDACXX_NOEXCEPT_RETURN(...) \
+    noexcept(noexcept(__VA_ARGS__))        \
+    {                                      \
+      return __VA_ARGS__;                  \
+    }
+
+template <class _Fn, class _Tuple, size_t... _Id>
+inline _LIBCUDACXX_INLINE_VISIBILITY constexpr decltype(auto)
+__apply_tuple_impl(_Fn&& __f, _Tuple&& __t, __tuple_indices<_Id...>) _LIBCUDACXX_NOEXCEPT_RETURN(
+    _CUDA_VSTD::__invoke(_CUDA_VSTD::forward<_Fn>(__f), _CUDA_VSTD::get<_Id>(_CUDA_VSTD::forward<_Tuple>(__t))...))
+
+    template <class _Fn, class _Tuple>
+    inline _LIBCUDACXX_INLINE_VISIBILITY constexpr decltype(auto) apply(_Fn&& __f, _Tuple&& __t)
+        _LIBCUDACXX_NOEXCEPT_RETURN(_CUDA_VSTD::__apply_tuple_impl(_CUDA_VSTD::forward<_Fn>(__f),
+            _CUDA_VSTD::forward<_Tuple>(__t), __make_tuple_indices_t<tuple_size_v<remove_reference_t<_Tuple>>>{}))
+
+            template <class _Tp, class _Tuple, size_t... _Idx>
+            inline _LIBCUDACXX_INLINE_VISIBILITY constexpr _Tp
+    __make_from_tuple_impl(_Tuple&& __t, __tuple_indices<_Idx...>)
+        _LIBCUDACXX_NOEXCEPT_RETURN(_Tp(_CUDA_VSTD::get<_Idx>(_CUDA_VSTD::forward<_Tuple>(__t))...))
+
+            template <class _Tp, class _Tuple>
+            inline _LIBCUDACXX_INLINE_VISIBILITY constexpr _Tp
+    make_from_tuple(_Tuple&& __t) _LIBCUDACXX_NOEXCEPT_RETURN(_CUDA_VSTD::__make_from_tuple_impl<_Tp>(
+        _CUDA_VSTD::forward<_Tuple>(__t), __make_tuple_indices_t<tuple_size_v<remove_reference_t<_Tuple>>>{}))
+
+#  undef _LIBCUDACXX_NOEXCEPT_RETURN
 
 #endif // _LIBCUDACXX_STD_VER > 14
 
-_LIBCUDACXX_END_NAMESPACE_STD
+        _LIBCUDACXX_END_NAMESPACE_STD
 
 #ifndef __cuda_std__
-#include <__pragma_pop>
+#  include <__pragma_pop>
 #endif //__cuda_std__
 
-#endif  // _LIBCUDACXX_TUPLE
+#endif // _LIBCUDACXX_TUPLE
diff --git a/libcudacxx/libcxx/docs/UsingLibcxx.rst b/libcudacxx/libcxx/docs/UsingLibcxx.rst
index 0e85967692..62f46779f8 100644
--- a/libcudacxx/libcxx/docs/UsingLibcxx.rst
+++ b/libcudacxx/libcxx/docs/UsingLibcxx.rst
@@ -165,33 +165,6 @@ thread safety annotations.
   headers. The intended use case is for clients who wish to use the libc++
   headers without taking a dependency on the libc++ library itself.
 
-**_LIBCUDACXX_ENABLE_TUPLE_IMPLICIT_REDUCED_ARITY_EXTENSION**:
-  This macro is used to re-enable an extension in `std::tuple` which allowed
-  it to be implicitly constructed from fewer initializers than contained
-  elements. Elements without an initializer are default constructed. For example:
-
-  .. code-block:: cpp
-
-    std::tuple<std::string, int, std::error_code> foo() {
-      return {"hello world", 42}; // default constructs error_code
-    }
-
-
-  Since libc++ 4.0 this extension has been disabled by default. This macro
-  may be defined to re-enable it in order to support existing code that depends
-  on the extension. New use of this extension should be discouraged.
-  See `PR 27374 <http://llvm.org/PR27374>`_ for more information.
-
-  Note: The "reduced-arity-initialization" extension is still offered but only
-  for explicit conversions. Example:
-
-  .. code-block:: cpp
-
-    auto foo() {
-      using Tup = std::tuple<std::string, int, std::error_code>;
-      return Tup{"hello world", 42}; // explicit constructor called. OK.
-    }
-
 **_LIBCUDACXX_DISABLE_ADDITIONAL_DIAGNOSTICS**:
   This macro disables the additional diagnostics generated by libc++ using the
   `diagnose_if` attribute. These additional diagnostics include checks for:
diff --git a/libcudacxx/libcxx/test/libcxx/utilities/tuple/tuple.tuple/tuple.cnstr/enable_reduced_arity_initialization_extension.pass.cpp b/libcudacxx/libcxx/test/libcxx/utilities/tuple/tuple.tuple/tuple.cnstr/enable_reduced_arity_initialization_extension.pass.cpp
deleted file mode 100644
index 14d4fcfbd6..0000000000
--- a/libcudacxx/libcxx/test/libcxx/utilities/tuple/tuple.tuple/tuple.cnstr/enable_reduced_arity_initialization_extension.pass.cpp
+++ /dev/null
@@ -1,119 +0,0 @@
-//===----------------------------------------------------------------------===//
-//
-// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
-// See https://llvm.org/LICENSE.txt for license information.
-// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
-//
-//===----------------------------------------------------------------------===//
-
-// <tuple>
-
-// template <class... Types> class tuple;
-
-// template <class... UTypes>
-//   explicit tuple(UTypes&&... u);
-
-// UNSUPPORTED: c++98, c++03
-
-// MODULES_DEFINES: _LIBCUDACXX_ENABLE_TUPLE_IMPLICIT_REDUCED_ARITY_EXTENSION
-#define _LIBCUDACXX_ENABLE_TUPLE_IMPLICIT_REDUCED_ARITY_EXTENSION
-#include <tuple>
-#include <cassert>
-#include <type_traits>
-#include <string>
-#include <system_error>
-
-#include "test_macros.h"
-#include "test_convertible.h"
-#include "MoveOnly.h"
-
-
-struct NoDefault { NoDefault() = delete; };
-
-
-// Make sure the _Up... constructor SFINAEs out when the types that
-// are not explicitly initialized are not all default constructible.
-// Otherwise, std::is_constructible would return true but instantiating
-// the constructor would fail.
-void test_default_constructible_extension_sfinae()
-{
-    typedef MoveOnly MO;
-    typedef NoDefault ND;
-    {
-        typedef std::tuple<MO, ND> Tuple;
-        static_assert(!std::is_constructible<Tuple, MO>::value, "");
-        static_assert(std::is_constructible<Tuple, MO, ND>::value, "");
-        static_assert(test_convertible<Tuple, MO, ND>(), "");
-    }
-    {
-        typedef std::tuple<MO, MO, ND> Tuple;
-        static_assert(!std::is_constructible<Tuple, MO, MO>::value, "");
-        static_assert(std::is_constructible<Tuple, MO, MO, ND>::value, "");
-        static_assert(test_convertible<Tuple, MO, MO, ND>(), "");
-    }
-    {
-        // Same idea as above but with a nested tuple type.
-        typedef std::tuple<MO, ND> Tuple;
-        typedef std::tuple<MO, Tuple, MO, MO> NestedTuple;
-
-        static_assert(!std::is_constructible<
-            NestedTuple, MO, MO, MO, MO>::value, "");
-        static_assert(std::is_constructible<
-            NestedTuple, MO, Tuple, MO, MO>::value, "");
-    }
-    {
-        typedef std::tuple<MO, int> Tuple;
-        typedef std::tuple<MO, Tuple, MO, MO> NestedTuple;
-
-        static_assert(std::is_constructible<
-            NestedTuple, MO, MO, MO, MO>::value, "");
-        static_assert(test_convertible<
-            NestedTuple, MO, MO, MO, MO>(), "");
-
-        static_assert(std::is_constructible<
-            NestedTuple, MO, Tuple, MO, MO>::value, "");
-        static_assert(test_convertible<
-            NestedTuple, MO, Tuple, MO, MO>(), "");
-    }
-}
-
-std::tuple<std::string, int, std::error_code> doc_example() {
-      return {"hello world", 42};
-}
-
-// Test that the example given in UsingLibcxx.rst actually works.
-void test_example_from_docs() {
-  auto tup = doc_example();
-  assert(std::get<0>(tup) == "hello world");
-  assert(std::get<1>(tup) == 42);
-  assert(std::get<2>(tup) == std::error_code{});
-}
-
-int main(int, char**)
-{
-
-    {
-        using E = MoveOnly;
-        using Tup = std::tuple<E, E, E>;
-        static_assert(test_convertible<Tup, E, E, E>(), "");
-
-        Tup t = {E(0), E(1)};
-        static_assert(test_convertible<Tup, E, E>(), "");
-        assert(std::get<0>(t) == 0);
-        assert(std::get<1>(t) == 1);
-        assert(std::get<2>(t) == MoveOnly());
-
-        Tup t2 = {E(0)};
-        static_assert(test_convertible<Tup, E>(), "");
-        assert(std::get<0>(t) == 0);
-        assert(std::get<1>(t) == E());
-        assert(std::get<2>(t) == E());
-        unused(t2);
-    }
-    // Check that SFINAE is properly applied with the default reduced arity
-    // constructor extensions.
-    test_default_constructible_extension_sfinae();
-    test_example_from_docs();
-
-  return 0;
-}