Merge branch 'main' into document-nvtarget-macros

libcudacxx/include/cuda/std/__cccl/builtin.h

@@ -288,6 +288,48 @@
 #  define _CCCL_BUILTIN_LINE() __LINE__
 #endif // _CCCL_CUDACC_BELOW(11, 3)
 
+#if _CCCL_CHECK_BUILTIN(builtin_huge_valf) || _CCCL_COMPILER(MSVC) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_HUGE_VALF() __builtin_huge_valf()
+#endif // _CCCL_CHECK_BUILTIN(builtin_huge_valf)
+
+#if _CCCL_CHECK_BUILTIN(builtin_huge_val) || _CCCL_COMPILER(MSVC) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_HUGE_VAL() __builtin_huge_val()
+#endif // _CCCL_CHECK_BUILTIN(builtin_huge_val)
+
+#if _CCCL_CHECK_BUILTIN(builtin_huge_vall) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_HUGE_VALL() __builtin_huge_vall()
+#elif _CCCL_COMPILER(MSVC)
+#  define _CCCL_BUILTIN_HUGE_VALL() static_cast<long double>(__builtin_huge_val())
+#endif // _CCCL_CHECK_BUILTIN(builtin_huge_vall)
+
+#if _CCCL_CHECK_BUILTIN(builtin_nanf) || _CCCL_COMPILER(MSVC) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_NANF(...) __builtin_nanf(__VA_ARGS__)
+#endif // _CCCL_CHECK_BUILTIN(builtin_nanf)
+
+#if _CCCL_CHECK_BUILTIN(builtin_nan) || _CCCL_COMPILER(MSVC) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_NAN(...) __builtin_nan(__VA_ARGS__)
+#endif // _CCCL_CHECK_BUILTIN(builtin_nan)
+
+#if _CCCL_CHECK_BUILTIN(builtin_nanl) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_NANL(...) __builtin_nanl(__VA_ARGS__)
+#elif _CCCL_COMPILER(MSVC)
+#  define _CCCL_BUILTIN_NANL(...) static_cast<long double>(__builtin_nan(__VA_ARGS__))
+#endif // _CCCL_CHECK_BUILTIN(builtin_nanl)
+
+#if _CCCL_CHECK_BUILTIN(builtin_nansf) || _CCCL_COMPILER(MSVC) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_NANSF(...) __builtin_nansf(__VA_ARGS__)
+#endif // _CCCL_CHECK_BUILTIN(builtin_nansf)
+
+#if _CCCL_CHECK_BUILTIN(builtin_nans) || _CCCL_COMPILER(MSVC) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_NANS(...) __builtin_nans(__VA_ARGS__)
+#endif // _CCCL_CHECK_BUILTIN(builtin_nans)
+
+#if _CCCL_CHECK_BUILTIN(builtin_nansl) || _CCCL_COMPILER(GCC, <, 10)
+#  define _CCCL_BUILTIN_NANSL(...) __builtin_nansl(__VA_ARGS__)
+#elif _CCCL_COMPILER(MSVC)
+#  define _CCCL_BUILTIN_NANSL(...) static_cast<long double>(__builtin_nans(__VA_ARGS__))
+#endif // _CCCL_CHECK_BUILTIN(builtin_nansl)
+
 #if _CCCL_CHECK_BUILTIN(builtin_log) || _CCCL_COMPILER(GCC)
 #  define _CCCL_BUILTIN_LOGF(...) __builtin_logf(__VA_ARGS__)
 #  define _CCCL_BUILTIN_LOG(...)  __builtin_log(__VA_ARGS__)

libcudacxx/include/cuda/std/__internal/namespaces.h

@@ -75,7 +75,7 @@
 #  define _CUDA_VSTD_NOVERSION ::cuda::std
 #  define _CUDA_VSTD           ::cuda::std::_LIBCUDACXX_ABI_NAMESPACE
 #  define _CUDA_VRANGES        ::cuda::std::ranges::_LIBCUDACXX_ABI_NAMESPACE
-#  define _CUDA_VIEWS          ::cuda::std::ranges::views::_LIBCUDACXX_CUDA_ABI_NAMESPACE
+#  define _CUDA_VIEWS          ::cuda::std::ranges::views::_LIBCUDACXX_ABI_NAMESPACE
 #  define _CUDA_VMR            ::cuda::mr::_LIBCUDACXX_ABI_NAMESPACE
 #  define _CUDA_VPTX           ::cuda::ptx::_LIBCUDACXX_ABI_NAMESPACE
 #  define _CUDA_VSTD_FS        ::cuda::std::__fs::filesystem::_LIBCUDACXX_ABI_NAMESPACE

libcudacxx/include/cuda/std/detail/libcxx/include/support/win32/limits_msvc_win32.h → libcudacxx/include/cuda/std/__limits/msvc_win32.h

@@ -1,14 +1,16 @@
 // -*- C++ -*-
-//===------------------ support/win32/limits_msvc_win32.h -----------------===//
+//===----------------------------------------------------------------------===//
 //
-// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
+// Part of libcu++, the C++ Standard Library for your entire system,
+// 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
+// SPDX-FileCopyrightText: Copyright (c) 2024 NVIDIA CORPORATION & AFFILIATES.
 //
 //===----------------------------------------------------------------------===//
 
-#ifndef _LIBCUDACXX_SUPPORT_WIN32_LIMITS_MSVC_WIN32_H
-#define _LIBCUDACXX_SUPPORT_WIN32_LIMITS_MSVC_WIN32_H
+#ifndef _LIBCUDACXX___LIMITS_MSVC_WIN32_H
+#define _LIBCUDACXX___LIMITS_MSVC_WIN32_H
 
 #if defined(__MINGW32__)
 #  error "This header complements the Microsoft C Runtime library, and should not be included otherwise."
@@ -19,8 +21,6 @@
 
 #include <float.h> // limit constants
 #include <limits.h> // CHAR_BIT
-#include <math.h> // HUGE_VAL
-#include <ymath.h> // internal MSVC header providing the needed functionality
 
 #define __CHAR_BIT__ CHAR_BIT
 
@@ -63,15 +63,4 @@
 // predefined by MinGW GCC
 #define __LDBL_DENORM_MIN__ 3.64519953188247460253e-4951L
 
-// __builtin replacements/workarounds
-#if _MSC_VER < 1934
-#  define __builtin_huge_vall()    _LInf._Long_double
-#  define __builtin_nanl(__dummmy) _LNan._Long_double
-#  define __builtin_nansl(__dummy) _LSnan._Long_double
-#else
-#  define __builtin_huge_vall() __builtin_huge_val()
-#  define __builtin_nanl(__v)   __builtin_nan(__v)
-#  define __builtin_nansl(__v)  __builtin_nans(__v)
-#endif
-
-#endif // _LIBCUDACXX_SUPPORT_WIN32_LIMITS_MSVC_WIN32_H
+#endif // _LIBCUDACXX___LIMITS_MSVC_WIN32_H

libcudacxx/include/cuda/std/__memory/allocator_traits.h

@@ -28,6 +28,8 @@
 #include <cuda/std/__type_traits/is_copy_constructible.h>
 #include <cuda/std/__type_traits/is_empty.h>
 #include <cuda/std/__type_traits/is_move_constructible.h>
+#include <cuda/std/__type_traits/is_same.h>
+#include <cuda/std/__type_traits/is_trivially_move_constructible.h>
 #include <cuda/std/__type_traits/make_unsigned.h>
 #include <cuda/std/__type_traits/remove_reference.h>
 #include <cuda/std/__type_traits/void_t.h>

libcudacxx/include/cuda/std/__memory/temporary_buffer.h

@@ -26,10 +26,11 @@
 #include <cuda/std/__iterator/iterator_traits.h>
 #include <cuda/std/__memory/addressof.h>
 #include <cuda/std/__new_>
+#include <cuda/std/__type_traits/alignment_of.h>
 #include <cuda/std/__utility/move.h>
 #include <cuda/std/__utility/pair.h>
 #include <cuda/std/cstddef>
-#include <cuda/std/detail/libcxx/include/limits>
+#include <cuda/std/limits>
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 

libcudacxx/include/cuda/std/__ranges/empty_view.h

@@ -0,0 +1,80 @@
+// -*- C++ -*-
+//===----------------------------------------------------------------------===//
+//
+// 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
+// SPDX-FileCopyrightText: Copyright (c) 2025 NVIDIA CORPORATION & AFFILIATES
+//
+//===----------------------------------------------------------------------===//
+#ifndef _LIBCUDACXX___RANGES_EMPTY_VIEW_H
+#define _LIBCUDACXX___RANGES_EMPTY_VIEW_H
+
+#include <cuda/std/detail/__config>
+
+#if defined(_CCCL_IMPLICIT_SYSTEM_HEADER_GCC)
+#  pragma GCC system_header
+#elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_CLANG)
+#  pragma clang system_header
+#elif defined(_CCCL_IMPLICIT_SYSTEM_HEADER_MSVC)
+#  pragma system_header
+#endif // no system header
+
+#include <cuda/std/__ranges/enable_borrowed_range.h>
+#include <cuda/std/__ranges/view_interface.h>
+#include <cuda/std/__type_traits/is_object.h>
+
+#if _CCCL_STD_VER >= 2017 && !_CCCL_COMPILER(MSVC2017)
+
+_LIBCUDACXX_BEGIN_NAMESPACE_RANGES
+_LIBCUDACXX_BEGIN_NAMESPACE_RANGES_ABI
+
+_CCCL_TEMPLATE(class _Tp)
+_CCCL_REQUIRES(_CCCL_TRAIT(is_object, _Tp))
+class empty_view : public view_interface<empty_view<_Tp>>
+{
+public:
+  _LIBCUDACXX_HIDE_FROM_ABI static constexpr _Tp* begin() noexcept
+  {
+    return nullptr;
+  }
+  _LIBCUDACXX_HIDE_FROM_ABI static constexpr _Tp* end() noexcept
+  {
+    return nullptr;
+  }
+  _LIBCUDACXX_HIDE_FROM_ABI static constexpr _Tp* data() noexcept
+  {
+    return nullptr;
+  }
+  _LIBCUDACXX_HIDE_FROM_ABI static constexpr size_t size() noexcept
+  {
+    return 0;
+  }
+  _LIBCUDACXX_HIDE_FROM_ABI static constexpr bool empty() noexcept
+  {
+    return true;
+  }
+};
+
+_LIBCUDACXX_END_NAMESPACE_RANGES_ABI
+
+template <class _Tp>
+_CCCL_INLINE_VAR constexpr bool enable_borrowed_range<empty_view<_Tp>> = true;
+
+_LIBCUDACXX_END_NAMESPACE_RANGES
+
+_LIBCUDACXX_BEGIN_NAMESPACE_VIEWS
+
+#  if _CCCL_COMPILER(MSVC)
+template <class _Tp>
+_CCCL_INLINE_VAR constexpr empty_view<_Tp> empty{};
+#  else // ^^^ _CCCL_COMPILER_MSVC ^^^ / vvv !_CCCL_COMPILER_MSVC vvv
+template <class _Tp>
+_CCCL_GLOBAL_CONSTANT empty_view<_Tp> empty{};
+#  endif // !_CCCL_COMPILER_MSVC
+
+_LIBCUDACXX_END_NAMESPACE_VIEWS
+
+#endif // _CCCL_STD_VER >= 2017 && !_CCCL_COMPILER_MSVC_2017
+
+#endif // _LIBCUDACXX___RANGES_EMPTY_VIEW_H

libcudacxx/include/cuda/std/__ranges/subrange.h

@@ -244,7 +244,7 @@ class _CCCL_TYPE_VISIBILITY_DEFAULT subrange : public view_interface<subrange<_I
 
 public:
 #  if !defined(_CCCL_NO_CONCEPTS)
-  subrange()
+  _CCCL_HIDE_FROM_ABI subrange()
     requires default_initializable<_Iter>
   = default;
 #  else // ^^^ !_CCCL_NO_CONCEPTS ^^^ / vvv _CCCL_NO_CONCEPTS vvv
@@ -294,14 +294,15 @@ class _CCCL_TYPE_VISIBILITY_DEFAULT subrange : public view_interface<subrange<_I
       : subrange(_CUDA_VRANGES::begin(__range), _CUDA_VRANGES::end(__range), __n)
   {}
 
-#  if (!_CCCL_COMPILER(GCC) || _CCCL_COMPILER(GCC, >=, 9))
+  // This often ICEs all of clang and old gcc when it encounteres a rvalue subrange in a pipe
+#  if !defined(_CCCL_NO_CONCEPTS)
   _CCCL_TEMPLATE(class _Pair)
   _CCCL_REQUIRES(__pair_like<_Pair> _CCCL_AND __subrange_to_pair<_Iter, _Sent, _Kind, _Pair>)
   _LIBCUDACXX_HIDE_FROM_ABI constexpr operator _Pair() const
   {
     return _Pair(__begin_, __end_);
   }
-#  endif // (!_CCCL_COMPILER(GCC) || _CCCL_COMPILER(GCC, >=, 9))
+#  endif // !_CCCL_NO_CONCEPTS
 
   _CCCL_TEMPLATE(class _It = _Iter)
   _CCCL_REQUIRES(copyable<_It>)