diff --git a/libcudacxx/include/cuda/std/__utility/pair.h b/libcudacxx/include/cuda/std/__utility/pair.h
index 9269a09c745..8b17881446a 100644
--- a/libcudacxx/include/cuda/std/__utility/pair.h
+++ b/libcudacxx/include/cuda/std/__utility/pair.h
@@ -283,18 +283,21 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
       : __base(__t1, __t2)
   {}
 
-  template <class _U1          = _T1,
-            class _U2          = _T2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<_U1, _U2>,
+  template <class _U1, class _U2>
+  using __pair_constructible = typename __pair_constraints<_T1, _T2>::template __constructible<_U1, _U2>;
+
+  template <class _U1                                                  = _T1,
+            class _U2                                                  = _T2,
+            class _Constraints                                         = __pair_constructible<_U1, _U2>,
             __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
   _LIBCUDACXX_INLINE_VISIBILITY explicit constexpr pair(_U1&& __u1, _U2&& __u2) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, _U1) && _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T2, _U2))
       : __base(_CUDA_VSTD::forward<_U1>(__u1), _CUDA_VSTD::forward<_U2>(__u2))
   {}
 
-  template <class _U1          = _T1,
-            class _U2          = _T2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<_U1, _U2>,
+  template <class _U1                                                  = _T1,
+            class _U2                                                  = _T2,
+            class _Constraints                                         = __pair_constructible<_U1, _U2>,
             __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
   _LIBCUDACXX_INLINE_VISIBILITY constexpr pair(_U1&& __u1, _U2&& __u2) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, _U1) && _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T2, _U2))
@@ -316,9 +319,9 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
   pair(pair const&) = default;
   pair(pair&&)      = default;
 
-  template <class _U1 = _T1,
-            class _U2 = _T2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<const _U1&, const _U2&>,
+  template <class _U1                                                  = _T1,
+            class _U2                                                  = _T2,
+            class _Constraints                                         = __pair_constructible<const _U1&, const _U2&>,
             __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
   _LIBCUDACXX_INLINE_VISIBILITY explicit _CCCL_CONSTEXPR_CXX14 pair(const pair<_U1, _U2>& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, const _U1&)
@@ -326,9 +329,9 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
       : __base(__p.first, __p.second)
   {}
 
-  template <class _U1 = _T1,
-            class _U2 = _T2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<const _U1&, const _U2&>,
+  template <class _U1                                                  = _T1,
+            class _U2                                                  = _T2,
+            class _Constraints                                         = __pair_constructible<const _U1&, const _U2&>,
             __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
   _LIBCUDACXX_INLINE_VISIBILITY _CCCL_CONSTEXPR_CXX14 pair(const pair<_U1, _U2>& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, const _U1&)
@@ -337,18 +340,18 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
   {}
 
   // move constructors
-  template <class _U1          = _T1,
-            class _U2          = _T2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<_U1, _U2>,
+  template <class _U1                                                  = _T1,
+            class _U2                                                  = _T2,
+            class _Constraints                                         = __pair_constructible<_U1, _U2>,
             __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
   _LIBCUDACXX_INLINE_VISIBILITY explicit _CCCL_CONSTEXPR_CXX14 pair(pair<_U1, _U2>&& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, _U1) && _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T2, _U2))
       : __base(_CUDA_VSTD::forward<_U1>(__p.first), _CUDA_VSTD::forward<_U2>(__p.second))
   {}
 
-  template <class _U1          = _T1,
-            class _U2          = _T2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<_U1, _U2>,
+  template <class _U1                                                  = _T1,
+            class _U2                                                  = _T2,
+            class _Constraints                                         = __pair_constructible<_U1, _U2>,
             __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
   _LIBCUDACXX_INLINE_VISIBILITY _CCCL_CONSTEXPR_CXX14 pair(pair<_U1, _U2>&& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, _U1) && _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T2, _U2))
@@ -359,7 +362,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
 #if defined(__cuda_std__) && !defined(_CCCL_COMPILER_NVRTC)
   template <class _U1,
             class _U2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<const _U1&, const _U2&>,
+            class _Constraints                                         = __pair_constructible<const _U1&, const _U2&>,
             __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
   _CCCL_HOST _LIBCUDACXX_HIDE_FROM_ABI explicit _CCCL_CONSTEXPR_CXX14 pair(const ::std::pair<_U1, _U2>& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, const _U1&)
@@ -369,7 +372,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
 
   template <class _U1,
             class _U2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<const _U1&, const _U2&>,
+            class _Constraints                                         = __pair_constructible<const _U1&, const _U2&>,
             __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
   _CCCL_HOST _LIBCUDACXX_HIDE_FROM_ABI _CCCL_CONSTEXPR_CXX14 pair(const ::std::pair<_U1, _U2>& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, const _U1&)
@@ -379,7 +382,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
 
   template <class _U1,
             class _U2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<_U1, _U2>,
+            class _Constraints                                         = __pair_constructible<_U1, _U2>,
             __enable_if_t<_Constraints::__explicit_constructible, int> = 0>
   _CCCL_HOST _LIBCUDACXX_HIDE_FROM_ABI explicit _CCCL_CONSTEXPR_CXX14 pair(::std::pair<_U1, _U2>&& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, _U1) && _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T2, _U2))
@@ -388,7 +391,7 @@ struct _LIBCUDACXX_TEMPLATE_VIS pair : public __pair_base<_T1, _T2>
 
   template <class _U1,
             class _U2,
-            class _Constraints = typename __pair_constraints<_T1, _T2>::template __constructible<_U1, _U2>,
+            class _Constraints                                         = __pair_constructible<_U1, _U2>,
             __enable_if_t<_Constraints::__implicit_constructible, int> = 0>
   _CCCL_HOST _LIBCUDACXX_HIDE_FROM_ABI _CCCL_CONSTEXPR_CXX14 pair(::std::pair<_U1, _U2>&& __p) noexcept(
     _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T1, _U1) && _LIBCUDACXX_TRAIT(is_nothrow_constructible, _T2, _U2))
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/tuple b/libcudacxx/include/cuda/std/detail/libcxx/include/tuple
index 198fa1eafd8..4b0bed399e6 100644
--- a/libcudacxx/include/cuda/std/detail/libcxx/include/tuple
+++ b/libcudacxx/include/cuda/std/detail/libcxx/include/tuple
@@ -194,10 +194,6 @@ template <class... Types>
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-template<class>
-struct  __is_tuple_of_iterator_references : false_type
-{};
-
 // __tuple_leaf
 struct __tuple_leaf_default_constructor_tag {};
 
@@ -852,15 +848,6 @@ public:
               _Tp...>::template __tuple_like_constraints<_Tuple>,
           __invalid_tuple_constraints>;
 
-  // Horrible hack to make tuple_of_iterator_references work
-  template <class _TupleOfIteratorReferences,
-            __enable_if_t<__is_tuple_of_iterator_references<_TupleOfIteratorReferences>::value, int> = 0,
-            __enable_if_t<(tuple_size<_TupleOfIteratorReferences>::value == sizeof...(_Tp)), int> = 0>
-  _LIBCUDACXX_INLINE_VISIBILITY _CCCL_CONSTEXPR_CXX14 tuple(_TupleOfIteratorReferences&& __t)
-      : tuple(_CUDA_VSTD::forward<_TupleOfIteratorReferences>(__t).template __to_tuple<_Tp...>(
-        __make_tuple_indices_t<sizeof...(_Tp)>()))
-  {}
-
   template <
       class _Tuple, class _Constraints = __tuple_like_constraints<_Tuple>,
       __enable_if_t<!_PackExpandsToThisTuple<_Tuple>::value, int> = 0,
diff --git a/thrust/testing/pair.cu b/thrust/testing/pair.cu
index 1f1ddcf4d2e..cab025444bd 100644
--- a/thrust/testing/pair.cu
+++ b/thrust/testing/pair.cu
@@ -322,4 +322,16 @@ void TestPairStructuredBindings(void)
   ASSERT_EQUAL(b, b2);
 }
 DECLARE_UNITTEST(TestPairStructuredBindings);
-#endif
+
+void TestPairCTAD(void)
+{
+  const int a = 42;
+  const int b = 1337;
+  thrust::pair p(a, b);
+
+  auto [a2, b2] = p;
+  ASSERT_EQUAL(a, a2);
+  ASSERT_EQUAL(b, b2);
+}
+DECLARE_UNITTEST(TestPairCTAD);
+#endif // _CCCL_STD_VER >= 2017
diff --git a/thrust/testing/tuple.cu b/thrust/testing/tuple.cu
index c3cb1b23bab..372b095d556 100644
--- a/thrust/testing/tuple.cu
+++ b/thrust/testing/tuple.cu
@@ -511,7 +511,23 @@ void TestTupleStructuredBindings(void)
   ASSERT_EQUAL(c, c2);
 }
 DECLARE_UNITTEST(TestTupleStructuredBindings);
-#endif
+
+#if !defined(_CCCL_COMPILER_MSVC_2017)
+void TestTupleCTAD(void)
+{
+  const int a   = 0;
+  const char b  = 42;
+  const short c = 1337;
+  thrust::tuple t(a, b, c);
+
+  auto [a2, b2, c2] = t;
+  ASSERT_EQUAL(a, a2);
+  ASSERT_EQUAL(b, b2);
+  ASSERT_EQUAL(c, c2);
+}
+DECLARE_UNITTEST(TestTupleCTAD);
+#endif // !_CCCL_COMPILER_MSVC_2017
+#endif // _CCCL_STD_VER >= 2017
 
 // Ensure that we are backwards compatible with the old thrust::tuple implementation
 static_assert(
diff --git a/thrust/thrust/detail/functional/actor.h b/thrust/thrust/detail/functional/actor.h
index 0dd0560e69e..3f30d0a1570 100644
--- a/thrust/thrust/detail/functional/actor.h
+++ b/thrust/thrust/detail/functional/actor.h
@@ -65,7 +65,7 @@ struct actor : Eval
 {
   typedef Eval eval_type;
 
-  _CCCL_HOST_DEVICE constexpr actor();
+  constexpr actor() = default;
 
   _CCCL_HOST_DEVICE actor(const Eval& base);
 
diff --git a/thrust/thrust/detail/functional/actor.inl b/thrust/thrust/detail/functional/actor.inl
index 81146a594c6..64d367ed15f 100644
--- a/thrust/thrust/detail/functional/actor.inl
+++ b/thrust/thrust/detail/functional/actor.inl
@@ -48,11 +48,6 @@ namespace detail
 namespace functional
 {
 
-template <typename Eval>
-_CCCL_HOST_DEVICE constexpr actor<Eval>::actor()
-    : eval_type()
-{}
-
 template <typename Eval>
 _CCCL_HOST_DEVICE actor<Eval>::actor(const Eval& base)
     : eval_type(base)
diff --git a/thrust/thrust/iterator/detail/tuple_of_iterator_references.h b/thrust/thrust/iterator/detail/tuple_of_iterator_references.h
index 574715e4a77..bb43c9d1c4b 100644
--- a/thrust/thrust/iterator/detail/tuple_of_iterator_references.h
+++ b/thrust/thrust/iterator/detail/tuple_of_iterator_references.h
@@ -67,9 +67,7 @@ class tuple_of_iterator_references : public thrust::tuple<Ts...>
   using super_t = thrust::tuple<Ts...>;
   using super_t::super_t;
 
-  inline _CCCL_HOST_DEVICE tuple_of_iterator_references()
-      : super_t()
-  {}
+  tuple_of_iterator_references() = default;
 
   // allow implicit construction from tuple<refs>
   inline _CCCL_HOST_DEVICE tuple_of_iterator_references(const super_t& other)
@@ -138,15 +136,15 @@ class tuple_of_iterator_references : public thrust::tuple<Ts...>
 
 } // namespace detail
 
-THRUST_NAMESPACE_END
-
-_LIBCUDACXX_BEGIN_NAMESPACE_STD
-
 template <class... Ts>
 struct __is_tuple_of_iterator_references<THRUST_NS_QUALIFIER::detail::tuple_of_iterator_references<Ts...>>
-    : integral_constant<bool, true>
+    : _CUDA_VSTD::true_type
 {};
 
+THRUST_NAMESPACE_END
+
+_LIBCUDACXX_BEGIN_NAMESPACE_STD
+
 // define tuple_size, tuple_element, etc.
 template <class... Ts>
 struct tuple_size<THRUST_NS_QUALIFIER::detail::tuple_of_iterator_references<Ts...>>
diff --git a/thrust/thrust/optional.h b/thrust/thrust/optional.h
index be186a9d78a..f7822324907 100644
--- a/thrust/thrust/optional.h
+++ b/thrust/thrust/optional.h
@@ -838,7 +838,12 @@ struct nullopt_t
 /// void foo (thrust::optional<int>);
 /// foo(thrust::nullopt); //pass an empty optional
 /// ```
-static constexpr nullopt_t nullopt{nullopt_t::do_not_use{}, nullopt_t::do_not_use{}};
+#ifdef __CUDA_ARCH__
+__device__ static _LIBCUDACXX_CONSTEXPR_GLOBAL
+#else
+static constexpr
+#endif // __CUDA_ARCH__
+  nullopt_t nullopt{nullopt_t::do_not_use{}, nullopt_t::do_not_use{}};
 
 class bad_optional_access : public std::exception
 {
diff --git a/thrust/thrust/pair.h b/thrust/thrust/pair.h
index 4fd2e8f8333..53d5cc93edf 100644
--- a/thrust/thrust/pair.h
+++ b/thrust/thrust/pair.h
@@ -50,7 +50,7 @@ THRUST_NAMESPACE_BEGIN
  *  \tparam T A \c pair type of interest.
  */
 template <size_t N, class T>
-using tuple_element = ::cuda::std::tuple_element<N, T>;
+using tuple_element = _CUDA_VSTD::tuple_element<N, T>;
 
 /*! This convenience metafunction is included for compatibility with
  *  \p tuple. It returns \c 2, the number of elements of a \p pair,
@@ -59,7 +59,7 @@ using tuple_element = ::cuda::std::tuple_element<N, T>;
  *  \tparam Pair A \c pair type of interest.
  */
 template <class T>
-using tuple_size = ::cuda::std::tuple_size<T>;
+using tuple_size = _CUDA_VSTD::tuple_size<T>;
 
 /*! \p pair is a generic data structure encapsulating a heterogeneous
  *  pair of values.
@@ -73,10 +73,49 @@ using tuple_size = ::cuda::std::tuple_size<T>;
  *          provided by <tt>pair::second_type</tt>.
  */
 template <class T, class U>
-using pair = ::cuda::std::pair<T, U>;
-
-using ::cuda::std::get;
-using ::cuda::std::make_pair;
+struct pair : public _CUDA_VSTD::pair<T, U>
+{
+  using super_t = _CUDA_VSTD::pair<T, U>;
+  using super_t::super_t;
+
+#if (defined(_CCCL_COMPILER_GCC) && __GNUC__ < 9) || (defined(_CCCL_COMPILER_CLANG) && __clang_major__ < 12)
+  // For whatever reason nvcc complains about that constructor being used before being defined in a constexpr variable
+  constexpr pair() = default;
+
+  template <class _U1          = T,
+            class _U2          = U,
+            class _Constraints = typename _CUDA_VSTD::__pair_constraints<T, U>::template __constructible<_U1, _U2>,
+            _CUDA_VSTD::__enable_if_t<_Constraints::__implicit_constructible, int> = 0>
+  _CCCL_HOST_DEVICE constexpr pair(_U1&& __u1, _U2&& __u2)
+      : super_t(_CUDA_VSTD::forward<_U1>(__u1), _CUDA_VSTD::forward<_U2>(__u2))
+  {}
+#endif // _CCCL_COMPILER_GCC < 9 || _CCCL_COMPILER_CLANG < 12
+};
+
+#if _CCCL_STD_VER >= 2017
+template <class _T1, class _T2>
+_CCCL_HOST_DEVICE pair(_T1, _T2) -> pair<_T1, _T2>;
+#endif // _CCCL_STD_VER >= 2017
+
+template <class T1, class T2>
+inline _CCCL_HOST_DEVICE
+  _CUDA_VSTD::__enable_if_t<_CUDA_VSTD::__is_swappable<T1>::value && _CUDA_VSTD::__is_swappable<T2>::value, void>
+  swap(pair<T1, T2>& lhs, pair<T1, T2>& rhs) noexcept(
+    (_CUDA_VSTD::__is_nothrow_swappable<T1>::value && _CUDA_VSTD::__is_nothrow_swappable<T2>::value))
+{
+  lhs.swap(rhs);
+}
+
+template <class T1, class T2>
+inline _CCCL_HOST_DEVICE
+  pair<typename _CUDA_VSTD::__unwrap_ref_decay<T1>::type, typename _CUDA_VSTD::__unwrap_ref_decay<T2>::type>
+  make_pair(T1&& t1, T2&& t2)
+{
+  return pair<typename _CUDA_VSTD::__unwrap_ref_decay<T1>::type, typename _CUDA_VSTD::__unwrap_ref_decay<T2>::type>(
+    _CUDA_VSTD::forward<T1>(t1), _CUDA_VSTD::forward<T2>(t2));
+}
+
+using _CUDA_VSTD::get;
 
 /*! \endcond
  */
@@ -88,3 +127,38 @@ using ::cuda::std::make_pair;
  */
 
 THRUST_NAMESPACE_END
+
+_LIBCUDACXX_BEGIN_NAMESPACE_STD
+
+template <class T1, class T2>
+struct tuple_size<THRUST_NS_QUALIFIER::pair<T1, T2>> : tuple_size<pair<T1, T2>>
+{};
+
+template <size_t Id, class T1, class T2>
+struct tuple_element<Id, THRUST_NS_QUALIFIER::pair<T1, T2>> : tuple_element<Id, pair<T1, T2>>
+{};
+
+template <class T1, class T2>
+struct __tuple_like_ext<THRUST_NS_QUALIFIER::pair<T1, T2>> : true_type
+{};
+
+_LIBCUDACXX_END_NAMESPACE_STD
+
+// This is a workaround for the fact that structured bindings require that the specializations of
+// `tuple_size` and `tuple_element` reside in namespace std (https://eel.is/c++draft/dcl.struct.bind#4).
+// See https://github.com/NVIDIA/libcudacxx/issues/316 for a short discussion
+#if _CCCL_STD_VER >= 2017
+
+#  include <utility>
+
+namespace std
+{
+template <class T1, class T2>
+struct tuple_size<THRUST_NS_QUALIFIER::pair<T1, T2>> : tuple_size<pair<T1, T2>>
+{};
+
+template <size_t Id, class T1, class T2>
+struct tuple_element<Id, THRUST_NS_QUALIFIER::pair<T1, T2>> : tuple_element<Id, pair<T1, T2>>
+{};
+} // namespace std
+#endif // _CCCL_STD_VER >= 2017
diff --git a/thrust/thrust/tuple.h b/thrust/thrust/tuple.h
index 5fa1d7f2f7c..0a00756e94a 100644
--- a/thrust/thrust/tuple.h
+++ b/thrust/thrust/tuple.h
@@ -94,28 +94,32 @@ _CCCL_HOST_DEVICE inline bool operator>(const null_type&, const null_type&)
  *  \tparam N This parameter selects the element of interest.
  *  \tparam T A \c tuple type of interest.
  *
- *  \see pair
+ *  \see tuple
  *  \see tuple
  */
 template <size_t N, class T>
-using tuple_element = ::cuda::std::tuple_element<N, T>;
+using tuple_element = _CUDA_VSTD::tuple_element<N, T>;
 
 /*! This metafunction returns the number of elements
  *  of a \p tuple type of interest.
  *
  *  \tparam T A \c tuple type of interest.
  *
- *  \see pair
+ *  \see tuple
  *  \see tuple
  */
 template <class T>
-using tuple_size = ::cuda::std::tuple_size<T>;
+using tuple_size = _CUDA_VSTD::tuple_size<T>;
+
+template <class>
+struct __is_tuple_of_iterator_references : _CUDA_VSTD::false_type
+{};
 
 /*! \brief \p tuple is a class template that can be instantiated with up to ten
  *  arguments. Each template argument specifies the type of element in the \p
  *  tuple. Consequently, tuples are heterogeneous, fixed-size collections of
  *  values. An instantiation of \p tuple with two arguments is similar to an
- *  instantiation of \p pair with the same two arguments. Individual elements
+ *  instantiation of \p tuple with the same two arguments. Individual elements
  *  of a \p tuple may be accessed with the \p get function.
  *
  *  \tparam TN The type of the <tt>N</tt> \c tuple element. Thrust's \p tuple
@@ -143,20 +147,86 @@ using tuple_size = ::cuda::std::tuple_size<T>;
  *  }
  *  \endcode
  *
- *  \see pair
+ *  \see tuple
  *  \see get
  *  \see make_tuple
  *  \see tuple_element
  *  \see tuple_size
  *  \see tie
  */
-template <class... T>
-using tuple = ::cuda::std::tuple<T...>;
+#if defined(_CCCL_COMPILER_MSVC_2017)
+// MSVC2017 either cannot find the constructors of the base class ::cuda::std::tuple or it declares them ambiguous.
+// Make CI happy at the expense of MSVC2017 not having working CTAD
+
+template <class... Ts>
+using tuple = ::cuda::std::tuple<Ts...>;
 
 using ::cuda::std::get;
 using ::cuda::std::make_tuple;
 using ::cuda::std::tie;
 
+#else // !_CCCL_COMPILER_MSVC_2017
+template <class... Ts>
+struct tuple : public _CUDA_VSTD::tuple<Ts...>
+{
+  using super_t = _CUDA_VSTD::tuple<Ts...>;
+  using super_t::super_t;
+
+  tuple() = default;
+
+  template <class _TupleOfIteratorReferences,
+            _CUDA_VSTD::__enable_if_t<__is_tuple_of_iterator_references<_TupleOfIteratorReferences>::value, int> = 0,
+            _CUDA_VSTD::__enable_if_t<(tuple_size<_TupleOfIteratorReferences>::value == sizeof...(Ts)), int>     = 0>
+  _CCCL_HOST_DEVICE tuple(_TupleOfIteratorReferences&& tup)
+      : tuple(_CUDA_VSTD::forward<_TupleOfIteratorReferences>(tup).template __to_tuple<Ts...>(
+        _CUDA_VSTD::__make_tuple_indices_t<sizeof...(Ts)>()))
+  {}
+
+  _CCCL_EXEC_CHECK_DISABLE
+  template <class TupleLike,
+            _CUDA_VSTD::__enable_if_t<_CUDA_VSTD::__tuple_assignable<TupleLike, super_t>::value, int> = 0>
+  _CCCL_HOST_DEVICE tuple& operator=(TupleLike&& other)
+  {
+    super_t::operator=(_CUDA_VSTD::forward<TupleLike>(other));
+    return *this;
+  }
+};
+
+#  if _CCCL_STD_VER >= 2017
+template <class... Ts>
+_CCCL_HOST_DEVICE tuple(Ts...) -> tuple<Ts...>;
+
+template <class T1, class T2>
+struct pair;
+
+template <class T1, class T2>
+_CCCL_HOST_DEVICE tuple(pair<T1, T2>) -> tuple<T1, T2>;
+#  endif // _CCCL_STD_VER >= 2017
+
+template <class... Ts>
+inline _CCCL_HOST_DEVICE
+  _CUDA_VSTD::__enable_if_t<_CUDA_VSTD::__all<_CUDA_VSTD::__is_swappable<Ts>::value...>::value, void>
+  swap(tuple<Ts...>& __x,
+       tuple<Ts...>& __y) noexcept((_CUDA_VSTD::__all<_CUDA_VSTD::__is_nothrow_swappable<Ts>::value...>::value))
+{
+  __x.swap(__y);
+}
+
+template <class... Ts>
+inline _CCCL_HOST_DEVICE tuple<typename _CUDA_VSTD::__unwrap_ref_decay<Ts>::type...> make_tuple(Ts&&... __t)
+{
+  return tuple<typename _CUDA_VSTD::__unwrap_ref_decay<Ts>::type...>(_CUDA_VSTD::forward<Ts>(__t)...);
+}
+
+template <class... Ts>
+inline _CCCL_HOST_DEVICE tuple<Ts&...> tie(Ts&... ts) noexcept
+{
+  return tuple<Ts&...>(ts...);
+}
+
+using _CUDA_VSTD::get;
+#endif // !_CCCL_COMPILER_MSVC_2017
+
 /*! \endcond
  */
 
@@ -170,6 +240,18 @@ THRUST_NAMESPACE_END
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
+template <class... Ts>
+struct tuple_size<THRUST_NS_QUALIFIER::tuple<Ts...>> : tuple_size<tuple<Ts...>>
+{};
+
+template <size_t Id, class... Ts>
+struct tuple_element<Id, THRUST_NS_QUALIFIER::tuple<Ts...>> : tuple_element<Id, tuple<Ts...>>
+{};
+
+template <class... Ts>
+struct __tuple_like_ext<THRUST_NS_QUALIFIER::tuple<Ts...>> : true_type
+{};
+
 template <>
 struct tuple_size<tuple<THRUST_NS_QUALIFIER::null_type,
                         THRUST_NS_QUALIFIER::null_type,
@@ -278,3 +360,19 @@ struct tuple_size<tuple<T0, T1, T2, T3, T4, T5, T6, T7, T8, THRUST_NS_QUALIFIER:
 {};
 
 _LIBCUDACXX_END_NAMESPACE_STD
+
+// This is a workaround for the fact that structured bindings require that the specializations of
+// `tuple_size` and `tuple_element` reside in namespace std (https://eel.is/c++draft/dcl.struct.bind#4).
+// See https://github.com/NVIDIA/libcudacxx/issues/316 for a short discussion
+#if _CCCL_STD_VER >= 2017
+namespace std
+{
+template <class... Ts>
+struct tuple_size<THRUST_NS_QUALIFIER::tuple<Ts...>> : tuple_size<tuple<Ts...>>
+{};
+
+template <size_t Id, class... Ts>
+struct tuple_element<Id, THRUST_NS_QUALIFIER::tuple<Ts...>> : tuple_element<Id, tuple<Ts...>>
+{};
+} // namespace std
+#endif // _CCCL_STD_VER >= 2017