Port as_const to be usable in device code

NVIDIA · Oct 11, 2022 · 7f097a4 · 7f097a4
1 parent 5e74406
commit 7f097a4
Show file tree

Hide file tree

Showing 8 changed files with 85 additions and 14 deletions.
diff --git a/.upstream-tests/test/std/containers/views/mdspan/mdspan.mdspan.cons/ctad_extents.pass.cpp b/.upstream-tests/test/std/containers/views/mdspan/mdspan.mdspan.cons/ctad_extents.pass.cpp
@@ -55,7 +55,6 @@ int main(int, char**)
         CHECK_MDSPAN(m,d);
     }
 
-#if 0 // as_const is missing
     // extents from std::span
     {
         cuda::std::array<int, 1> d{42};
@@ -64,7 +63,6 @@ int main(int, char**)
 
         CHECK_MDSPAN(m,d);
     }
-#endif
 
     return 0;
 }
diff --git a/.upstream-tests/test/std/containers/views/mdspan/mdspan.mdspan.cons/extents.pass.cpp b/.upstream-tests/test/std/containers/views/mdspan/mdspan.mdspan.cons/extents.pass.cpp
@@ -56,7 +56,6 @@ int main(int, char**)
         CHECK_MDSPAN(m,d);
     }
 
-#if 0 // as_const is missing
     // extents from cuda::std::span
     {
         cuda::std::array<int, 1> d{42};
@@ -65,7 +64,6 @@ int main(int, char**)
 
         CHECK_MDSPAN(m,d);
     }
-#endif
 
     return 0;
 }

diff --git a/.upstream-tests/test/std/containers/views/mdspan/mdspan.mdspan.members/brackets_op.pass.cpp b/.upstream-tests/test/std/containers/views/mdspan/mdspan.mdspan.members/brackets_op.pass.cpp
@@ -31,13 +31,11 @@ int main(int, char**)
         assert( m(cuda::std::array<int,2>{1,0}) == 44 );
         assert( m(cuda::std::array<int,2>{1,1}) == 45 );
 
-#if 0 // missing as_const
         // span of indices
         assert( m(cuda::std::span{cuda::std::array<int,2>{0,0}}) == 42 );
         assert( m(cuda::std::span{cuda::std::array<int,2>{0,1}}) == 43 );
         assert( m(cuda::std::span{cuda::std::array<int,2>{1,0}}) == 44 );
         assert( m(cuda::std::span{cuda::std::array<int,2>{1,1}}) == 45 );
-#endif
     }
 
     return 0;

diff --git a/.upstream-tests/test/std/utilities/utility/as_const/as_const.fail.cpp b/.upstream-tests/test/std/utilities/utility/as_const/as_const.fail.cpp
@@ -0,0 +1,23 @@
+//===----------------------------------------------------------------------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: c++98, c++03, c++11
+
+// template <class T> constexpr add_const<T>& as_const(T& t) noexcept;      // C++17
+// template <class T>           add_const<T>& as_const(const T&&) = delete; // C++17
+
+#include <cuda/std/utility>
+
+struct S {int i;};
+
+int main(int, char**)
+{
+    cuda::std::as_const(S{});
+
+  return 0;
+}
diff --git a/.upstream-tests/test/std/utilities/utility/as_const/as_const.pass.cpp b/.upstream-tests/test/std/utilities/utility/as_const/as_const.pass.cpp
@@ -0,0 +1,52 @@
+//===----------------------------------------------------------------------===//
+//
+// 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
+//
+//===----------------------------------------------------------------------===//
+
+// UNSUPPORTED: c++98, c++03, c++11
+
+// template <class T> constexpr add_const<T>& as_const(T& t) noexcept;      // C++17
+// template <class T>           add_const<T>& as_const(const T&&) = delete; // C++17
+
+#include <cuda/std/cassert>
+#include <cuda/std/utility>
+
+#include "test_macros.h"
+
+struct S {int i;};
+__host__ __device__
+bool operator==(const S& x, const S& y) { return x.i == y.i; }
+__host__ __device__
+bool operator==(const volatile S& x, const volatile S& y) { return x.i == y.i; }
+
+template<typename T>
+__host__ __device__
+void test(T& t)
+{
+    static_assert(cuda::std::is_const<typename cuda::std::remove_reference<decltype(cuda::std::as_const                  (t))>::type>::value, "");
+    static_assert(cuda::std::is_const<typename cuda::std::remove_reference<decltype(cuda::std::as_const<               T>(t))>::type>::value, "");
+    static_assert(cuda::std::is_const<typename cuda::std::remove_reference<decltype(cuda::std::as_const<const          T>(t))>::type>::value, "");
+    static_assert(cuda::std::is_const<typename cuda::std::remove_reference<decltype(cuda::std::as_const<volatile       T>(t))>::type>::value, "");
+    static_assert(cuda::std::is_const<typename cuda::std::remove_reference<decltype(cuda::std::as_const<const volatile T>(t))>::type>::value, "");
+
+    assert(cuda::std::as_const(t) == t);
+    assert(cuda::std::as_const<               T>(t) == t);
+    assert(cuda::std::as_const<const          T>(t) == t);
+    assert(cuda::std::as_const<volatile       T>(t) == t);
+    assert(cuda::std::as_const<const volatile T>(t) == t);
+}
+
+int main(int, char**)
+{
+    int i = 3;
+    double d = 4.0;
+    S s{2};
+    test(i);
+    test(d);
+    test(s);
+
+  return 0;
+}
diff --git a/include/cuda/std/detail/libcxx/include/__mdspan/mdspan.hpp b/include/cuda/std/detail/libcxx/include/__mdspan/mdspan.hpp
@@ -187,7 +187,7 @@ class mdspan
   __MDSPAN_CONDITIONAL_EXPLICIT(_Np != rank_dynamic())
   __MDSPAN_INLINE_FUNCTION
   constexpr mdspan(data_handle_type __p, _CUDA_VSTD::span<_SizeType, _Np> __dynamic_extents)
-    : __members(std::move(__p), __map_acc_pair_t(mapping_type(extents_type(as_const(__dynamic_extents))), accessor_type()))
+    : __members(std::move(__p), __map_acc_pair_t(mapping_type(extents_type(_CUDA_VSTD::as_const(__dynamic_extents))), accessor_type()))
   { }
 #endif
 

diff --git a/include/cuda/std/detail/libcxx/include/numeric b/include/cuda/std/detail/libcxx/include/numeric
@@ -152,12 +152,12 @@ floating_point  midpoint(floating_point a, floating_point b);    // C++20
 #include <functional>
 #include <cmath> // for isnormal
 #include <version>
+#include <__pragma_push>
 
 #if defined(_LIBCUDACXX_USE_PRAGMA_GCC_SYSTEM_HEADER)
 #pragma GCC system_header
 #endif
 
-_LIBCUDACXX_PUSH_MACROS
 #include <__undef_macros>
 #endif // __cuda_std__
 
@@ -594,10 +594,12 @@ midpoint(_Fp __a, _Fp __b) noexcept
 
 _LIBCUDACXX_END_NAMESPACE_STD
 
-_LIBCUDACXX_POP_MACROS
-
 #if defined(_LIBCUDACXX_HAS_PARALLEL_ALGORITHMS) && _LIBCUDACXX_STD_VER >= 17
 #   include <__pstl_numeric>
 #endif
 
+#ifndef __cuda_std__
+#include <__pragma_pop>
+#endif //__cuda_std__
+
 #endif  // _LIBCUDACXX_NUMERIC
diff --git a/include/cuda/std/detail/libcxx/include/utility b/include/cuda/std/detail/libcxx/include/utility
@@ -278,13 +278,13 @@ move_if_noexcept(_Tp& __x) _NOEXCEPT
 {
     return _CUDA_VSTD::move(__x);
 }
+#endif //__cuda_std__
 
-#if _LIBCUDACXX_STD_VER > 14
-template <class _Tp> constexpr add_const_t<_Tp>& as_const(_Tp& __t) noexcept { return __t; }
-template <class _Tp>                        void as_const(const _Tp&&) = delete;
+#if _LIBCUDACXX_STD_VER > 11
+template <class _Tp> _LIBCUDACXX_INLINE_VISIBILITY constexpr add_const_t<_Tp>& as_const(_Tp& __t) noexcept { return __t; }
+template <class _Tp> _LIBCUDACXX_INLINE_VISIBILITY                        void as_const(const _Tp&&) = delete;
 #endif
 
-#endif //__cuda_std__
 
 struct _LIBCUDACXX_TEMPLATE_VIS piecewise_construct_t { explicit piecewise_construct_t() = default; };
 #if defined(_LIBCUDACXX_CXX03_LANG) || defined(_LIBCUDACXX_BUILDING_LIBRARY)