Unify handling of dialects (#1200)

* Unify handling of dialects We currently have 3 hacky ways of detecting standard versions. To unify this we define a single `_CCCL_STD_VER` macro that works everywhere. For backward compatability we move libcu++ to a four digit counter and also keep the publicly available `CUB_CPP_DIALECT` and `THRUST_CPP_DIALECT` for user consumption. That said, we should never ever rely on the user to define those and always use our internal detection logic. Anything else is just plain broken.
NVIDIA · Dec 13, 2023 · fddf2fe · fddf2fe
1 parent 0ea74c4
commit fddf2fe
Show file tree

Hide file tree

Showing 2,654 changed files with 5,690 additions and 5,734 deletions.
diff --git a/cub/cub/detail/cpp_compatibility.cuh b/cub/cub/detail/cpp_compatibility.cuh
@@ -27,7 +27,7 @@
 #  pragma system_header
 #endif // no system header
 
-#if CUB_CPP_DIALECT >= 2017 && __cpp_if_constexpr
+#if _CCCL_STD_VER >= 2017 && __cpp_if_constexpr
 #  define CUB_IF_CONSTEXPR if constexpr
 #  define CUB_ELSE_IF_CONSTEXPR else if constexpr
 #else

diff --git a/cub/cub/detail/type_traits.cuh b/cub/cub/detail/type_traits.cuh
@@ -53,7 +53,7 @@ namespace detail {
 
 template <typename Invokable, typename... Args>
 using invoke_result_t =
-#if CUB_CPP_DIALECT < 2017
+#if _CCCL_STD_VER < 2017
   typename ::cuda::std::result_of<Invokable(Args...)>::type;
 #else // 2017+
   ::cuda::std::invoke_result_t<Invokable, Args...>;

diff --git a/cub/cub/thread/thread_operators.cuh b/cub/cub/thread/thread_operators.cuh
@@ -78,7 +78,7 @@ struct InequalityWrapper
   }
 };
 
-#if CUB_CPP_DIALECT > 2011
+#if _CCCL_STD_VER > 2011
 using Equality = ::cuda::std::equal_to<>;
 using Inequality = ::cuda::std::not_equal_to<>;
 using Sum = ::cuda::std::plus<>;

diff --git a/cub/cub/util_cpp_dialect.cuh b/cub/cub/util_cpp_dialect.cuh
@@ -41,7 +41,7 @@
 
 #include <cub/util_compiler.cuh>
 
-#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document 
+#ifndef DOXYGEN_SHOULD_SKIP_THIS // Do not document
 
 // Deprecation warnings may be silenced by defining the following macros. These
 // may be combined.
@@ -73,38 +73,7 @@
 #  define CUB_IGNORE_DEPRECATED_COMPILER
 #endif
 
-// Define this to override the built-in detection.
-#ifndef CUB_CPP_DIALECT
-
-// MSVC does not define __cplusplus correctly. _MSVC_LANG is used instead.
-// This macro is only defined in MSVC 2015U3+.
-#  ifdef _MSVC_LANG // Do not replace with CUB_HOST_COMPILER test (see above)
-// MSVC2015 reports C++14 but lacks extended constexpr support. Treat as C++11.
-#    if CUB_MSVC_VERSION < 1910 && _MSVC_LANG > 201103L /* MSVC < 2017 && CPP > 2011 */
-#      define CUB_CPLUSPLUS 201103L /* Fix to 2011 */
-#    else
-#      define CUB_CPLUSPLUS _MSVC_LANG /* We'll trust this for now. */
-#    endif // MSVC 2015 C++14 fix
-#  else
-#    define CUB_CPLUSPLUS __cplusplus
-#  endif
-
-// Detect current dialect:
-#  if CUB_CPLUSPLUS < 201103L
-#    define CUB_CPP_DIALECT 2003
-#  elif CUB_CPLUSPLUS < 201402L
-#    define CUB_CPP_DIALECT 2011
-#  elif CUB_CPLUSPLUS < 201703L
-#    define CUB_CPP_DIALECT 2014
-#  elif CUB_CPLUSPLUS == 201703L
-#    define CUB_CPP_DIALECT 2017
-#  elif CUB_CPLUSPLUS > 201703L // unknown, but is higher than 2017.
-#    define CUB_CPP_DIALECT 2020
-#  endif
-
-#  undef CUB_CPLUSPLUS // cleanup
-
-#endif // !CUB_CPP_DIALECT
+#define CUB_CPP_DIALECT _CCCL_STD_VER
 
 // Define CUB_COMPILER_DEPRECATION macro:
 #if CUB_HOST_COMPILER == CUB_HOST_COMPILER_MSVC
@@ -144,10 +113,10 @@
 #ifndef CUB_IGNORE_DEPRECATED_DIALECT
 
 // Dialect checks:
-#  if CUB_CPP_DIALECT < 2011
+#  if _CCCL_STD_VER < 2011
      // <C++11. Hard upgrade message:
      CUB_COMPILER_DEPRECATION(C++14);
-#  elif CUB_CPP_DIALECT == 2011 && !defined(CUB_IGNORE_DEPRECATED_CPP_11)
+#  elif _CCCL_STD_VER == 2011 && !defined(CUB_IGNORE_DEPRECATED_CPP_11)
      // =C++11. Soft upgrade message:
      CUB_COMPILER_DEPRECATION_SOFT(C++14, C++11);
 #  endif

diff --git a/cub/cub/util_deprecated.cuh b/cub/cub/util_deprecated.cuh
@@ -52,7 +52,7 @@
 #ifdef CUB_IGNORE_DEPRECATED_API
 #  define CUB_DEPRECATED
 #  define CUB_DEPRECATED_BECAUSE(MSG)
-#elif CUB_CPP_DIALECT >= 2014
+#elif _CCCL_STD_VER >= 2014
 #  define CUB_DEPRECATED [[deprecated]]
 #  define CUB_DEPRECATED_BECAUSE(MSG) [[deprecated(MSG)]]
 #elif CUB_HOST_COMPILER == CUB_HOST_COMPILER_MSVC

diff --git a/cub/docs/repo.toml b/cub/docs/repo.toml
@@ -84,6 +84,7 @@ doxygen_aliases = [
 # more information on the format can be found at:
 #     https://www.doxygen.nl/manual/config.html#cfg_predefined
 doxygen_predefined = [
+    "_CCCL_STD_VER",
     "CUB_DISABLE_NAMESPACE_MAGIC",
     "CUB_IGNORE_NAMESPACE_MAGIC_ERROR",
     "CUB_NAMESPACE_BEGIN=namespace cub {",

diff --git a/cub/examples/block/example_block_reduce_dyn_smem.cu b/cub/examples/block/example_block_reduce_dyn_smem.cu
@@ -47,7 +47,7 @@
 #include "../../test/test_util.h"
 
 // Some implementation details rely on c++14
-#if CUB_CPP_DIALECT >= 2014
+#if _CCCL_STD_VER >= 2014
 
 using namespace cub;
 
@@ -86,7 +86,7 @@ __global__ void BlockReduceKernel(
 
     // shared memory byte-array
     extern __shared__ __align__(alignof(ShmemLayout)) char smem[];
-    
+
     // cast to lvalue reference of expected type
     auto& temp_storage = reinterpret_cast<TempStorageT&>(smem);
 

diff --git a/cub/test/test_warning_suppression.cuh b/cub/test/test_warning_suppression.cuh
@@ -34,7 +34,7 @@
 // This can be fixed with `if constexpr` when available, but there's no way to
 // silence these pre-C++17.
 #if CUB_HOST_COMPILER == CUB_HOST_COMPILER_MSVC
-#  if CUB_CPP_DIALECT < 2017
+#  if _CCCL_STD_VER < 2017
 #    pragma warning(disable:4127)
 #  endif
 #endif
diff --git a/libcudacxx/include/cuda/memory_resource b/libcudacxx/include/cuda/memory_resource
@@ -91,7 +91,7 @@ class resource_ref {
 
 #include <cuda/std/detail/__pragma_push>
 
-#if _LIBCUDACXX_STD_VER > 11
+#if _CCCL_STD_VER > 2011
 _LIBCUDACXX_BEGIN_NAMESPACE_CUDA
 
 ///////////////////////////////////////////////////////////////////////////////
@@ -233,7 +233,7 @@ _LIBCUDACXX_CONCEPT async_resource = _LIBCUDACXX_FRAGMENT(__async_resource_, _Re
 /// \concept resource_with
 /// \brief The \c resource_with concept
 template <class _Resource, class... _Properties>
-#if _LIBCUDACXX_STD_VER < 17
+#if _CCCL_STD_VER < 2017
 _LIBCUDACXX_CONCEPT resource_with =
   resource<_Resource>&& _CUDA_VSTD::conjunction_v<_CUDA_VSTD::bool_constant<has_property<_Resource, _Properties>>...>;
 #else
@@ -243,7 +243,7 @@ _LIBCUDACXX_CONCEPT resource_with = resource<_Resource> && (has_property<_Resour
 /// \concept async_resource_with
 /// \brief The \c async_resource_with concept
 template <class _Resource, class... _Properties>
-#if _LIBCUDACXX_STD_VER < 17
+#if _CCCL_STD_VER < 2017
 _LIBCUDACXX_CONCEPT async_resource_with = async_resource<_Resource> &&
   _CUDA_VSTD::conjunction_v<_CUDA_VSTD::bool_constant<has_property<_Resource, _Properties>>...>;
 #else
@@ -530,7 +530,7 @@ public:
         , _Filtered_vtable<_Properties...>(_Filtered_vtable<_Properties...>::template _Create<_Resource>())
     {}
 
-    #if _LIBCUDACXX_STD_VER > 14
+    #if _CCCL_STD_VER > 2014
     _LIBCUDACXX_TEMPLATE(class... _OtherProperties)
       _LIBCUDACXX_REQUIRES( (_CUDA_VSTD::_One_of<_Properties, _OtherProperties...> && ...))
     #else
@@ -544,7 +544,7 @@ public:
         , _Filtered_vtable<_Properties...>(__ref)
     {}
 
-    #if _LIBCUDACXX_STD_VER > 14
+    #if _CCCL_STD_VER > 2014
     _LIBCUDACXX_TEMPLATE(class... _OtherProperties)
       _LIBCUDACXX_REQUIRES( (_Alloc_type == _AllocType::_Default)
         && (_CUDA_VSTD::_One_of<_Properties, _OtherProperties...> && ...))
@@ -560,7 +560,7 @@ public:
         , _Filtered_vtable<_Properties...>(__ref)
     {}
 
-    #if _LIBCUDACXX_STD_VER > 14
+    #if _CCCL_STD_VER > 2014
     _LIBCUDACXX_TEMPLATE(class... _OtherProperties)
       _LIBCUDACXX_REQUIRES((sizeof...(_Properties) == sizeof...(_OtherProperties))
             && (_CUDA_VSTD::_One_of<_Properties, _OtherProperties...> && ...))
@@ -576,7 +576,7 @@ public:
             && this->__static_vtable->__equal_fn(this->__object, __right.__object);
     }
 
-    #if _LIBCUDACXX_STD_VER > 14
+    #if _CCCL_STD_VER > 2014
     _LIBCUDACXX_TEMPLATE(class... _OtherProperties)
       _LIBCUDACXX_REQUIRES( (sizeof...(_Properties) == sizeof...(_OtherProperties))
             && (_CUDA_VSTD::_One_of<_Properties, _OtherProperties...> && ...))
@@ -623,7 +623,7 @@ struct host_accessible{};
 
 } // namespace mr
 _LIBCUDACXX_END_NAMESPACE_CUDA
-#endif // _LIBCUDACXX_STD_VER > 11
+#endif // _CCCL_STD_VER > 2011
 
 #include <cuda/std/detail/__pragma_pop>
 

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/CMakeLists.txt b/libcudacxx/include/cuda/std/detail/libcxx/include/CMakeLists.txt
@@ -48,6 +48,9 @@ set(files
   __bsd_locale_fallbacks.h
   __errc
   __cccl_config
+  __cccl/dialect.h
+  __cccl/version.h
+  __cccl/visibility.h
   __concepts/__concept_macros.h
   __concepts/_One_of.h
   __concepts/arithmetic.h

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/copy.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/copy.h
@@ -48,13 +48,13 @@ template <class _Tp, class _Up>
 inline _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX17 bool
 __dispatch_memmove(_Up* __result, _Tp* __first, const size_t __n)
 {
-#if _LIBCUDACXX_STD_VER >= 20
+#if _CCCL_STD_VER >= 2020
   if (_CUDA_VSTD::is_constant_evaluated())
   {
     return false;
   }
   else
-#endif // _LIBCUDACXX_STD_VER >= 20
+#endif // _CCCL_STD_VER >= 2020
   {
     // For now, we only ever use memmove on host
     // clang-format off

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/equal.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/equal.h
@@ -49,7 +49,7 @@ equal(_InputIterator1 __first1, _InputIterator1 __last1, _InputIterator2 __first
   return _CUDA_VSTD::equal(__first1, __last1, __first2, __equal_to{});
 }
 
-#if _LIBCUDACXX_STD_VER > 11
+#if _CCCL_STD_VER > 2011
 template <class _BinaryPredicate, class _InputIterator1, class _InputIterator2>
 _LIBCUDACXX_NODISCARD_EXT inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 __equal(_InputIterator1 __first1,
@@ -119,7 +119,7 @@ equal(_InputIterator1 __first1, _InputIterator1 __last1, _InputIterator2 __first
     typename iterator_traits<_InputIterator1>::iterator_category(),
     typename iterator_traits<_InputIterator2>::iterator_category());
 }
-#endif // _LIBCUDACXX_STD_VER > 11
+#endif // _CCCL_STD_VER > 2011
 
 _LIBCUDACXX_END_NAMESPACE_STD
 

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/is_permutation.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/is_permutation.h
@@ -108,7 +108,7 @@ is_permutation(_ForwardIterator1 __first1, _ForwardIterator1 __last1, _ForwardIt
   return _CUDA_VSTD::is_permutation(__first1, __last1, __first2, __equal_to{});
 }
 
-#if _LIBCUDACXX_STD_VER > 11
+#if _CCCL_STD_VER > 2011
 template <class _BinaryPredicate, class _ForwardIterator1, class _ForwardIterator2>
 _LIBCUDACXX_NODISCARD_EXT inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11 bool
 __is_permutation(
@@ -245,7 +245,7 @@ is_permutation(
     __iterator_category_type<_ForwardIterator1>{},
     __iterator_category_type<_ForwardIterator2>{});
 }
-#endif // _LIBCUDACXX_STD_VER > 11
+#endif // _CCCL_STD_VER > 2011
 
 _LIBCUDACXX_END_NAMESPACE_STD
 

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/mismatch.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/mismatch.h
@@ -50,7 +50,7 @@ _LIBCUDACXX_NODISCARD_EXT inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VI
   return _CUDA_VSTD::mismatch(__first1, __last1, __first2, __equal_to{});
 }
 
-#if _LIBCUDACXX_STD_VER > 11
+#if _CCCL_STD_VER > 2011
 template <class _InputIterator1, class _InputIterator2, class _BinaryPredicate>
 _LIBCUDACXX_NODISCARD_EXT inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
   pair<_InputIterator1, _InputIterator2>
@@ -77,7 +77,7 @@ _LIBCUDACXX_NODISCARD_EXT inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VI
 {
   return _CUDA_VSTD::mismatch(__first1, __last1, __first2, __last2, __equal_to{});
 }
-#endif // _LIBCUDACXX_STD_VER > 11
+#endif // _CCCL_STD_VER > 2011
 
 _LIBCUDACXX_END_NAMESPACE_STD
 

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/search.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__algorithm/search.h
@@ -172,7 +172,7 @@ _LIBCUDACXX_NODISCARD_EXT inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VI
   return _CUDA_VSTD::search(__first1, __last1, __first2, __last2, __equal_to{});
 }
 
-#if _LIBCUDACXX_STD_VER > 14
+#if _CCCL_STD_VER > 2014
 template <class _ForwardIterator, class _Searcher>
 _LIBCUDACXX_NODISCARD_EXT inline _LIBCUDACXX_HIDE_FROM_ABI _LIBCUDACXX_INLINE_VISIBILITY _LIBCUDACXX_CONSTEXPR_AFTER_CXX11
   _ForwardIterator

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__bit_reference b/libcudacxx/include/cuda/std/detail/libcxx/include/__bit_reference
@@ -1113,7 +1113,7 @@ private:
 
 public:
     _LIBCUDACXX_INLINE_VISIBILITY __bit_iterator() noexcept
-#if _LIBCUDACXX_STD_VER > 11
+#if _CCCL_STD_VER > 2011
     : __seg_(nullptr), __ctz_(0)
 #endif
     {}

diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl/dialect.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl/dialect.h
@@ -0,0 +1,44 @@
+//===----------------------------------------------------------------------===//
+//
+// 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) 2023 NVIDIA CORPORATION & AFFILIATES.
+//
+//===----------------------------------------------------------------------===//
+
+#ifndef __CCCL_DIALECT_H
+#define __CCCL_DIALECT_H
+
+#if defined(_CCCL_COMPILER_MSVC)
+#  if _MSVC_LANG <= 201103L
+#    define _CCCL_STD_VER 2011
+#  elif _MSVC_LANG <= 201402L
+#    define _CCCL_STD_VER 2014
+#  elif _MSVC_LANG <= 201703L
+#    define _CCCL_STD_VER 2017
+#  elif _MSVC_LANG <= 202002L
+#    define _CCCL_STD_VER 2020
+#  else
+#    define _CCCL_STD_VER 2023 // current year, or date of c++2b ratification
+#  endif
+#else // ^^^ _CCCL_COMPILER_MSVC ^^^ / vvv !_CCCL_COMPILER_MSVC vvv
+#  if __cplusplus <= 199711L
+#    define _CCCL_STD_VER 2003
+#  elif __cplusplus <= 201103L
+#    define _CCCL_STD_VER 2011
+#  elif __cplusplus <= 201402L
+#    define _CCCL_STD_VER 2014
+#  elif __cplusplus <= 201703L
+#    define _CCCL_STD_VER 2017
+#  elif __cplusplus <= 202002L
+#    define _CCCL_STD_VER 2020
+#  elif __cplusplus <= 202302L
+#    define _CCCL_STD_VER 2023
+#  else
+#    define _CCCL_STD_VER 2024 // current year, or date of c++2c ratification
+#  endif
+#endif // !_CCCL_COMPILER_MSVC
+
+#endif // __CCCL_DIALECT_H
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl_config b/libcudacxx/include/cuda/std/detail/libcxx/include/__cccl_config
@@ -165,6 +165,7 @@
 #  define _CCCL_NV_DIAG_DEFAULT(_WARNING)
 #endif // other compilers
 
+#include "__cccl/dialect.h"
 #include "__cccl/visibility.h"
 
 #endif // __CCCL_CONFIG
diff --git a/libcudacxx/include/cuda/std/detail/libcxx/include/__concepts/_One_of.h b/libcudacxx/include/cuda/std/detail/libcxx/include/__concepts/_One_of.h
@@ -28,13 +28,13 @@
 
 _LIBCUDACXX_BEGIN_NAMESPACE_STD
 
-#if _LIBCUDACXX_STD_VER > 14
+#if _CCCL_STD_VER > 2014
 template <class _Ty, class... _Others>
 _LIBCUDACXX_CONCEPT _One_of = (is_same_v<_Ty, _Others> || ...);
-#elif _LIBCUDACXX_STD_VER > 11
+#elif _CCCL_STD_VER > 2011
 template <class _Ty, class... _Others>
 _LIBCUDACXX_CONCEPT _One_of = _LIBCUDACXX_TRAIT(disjunction, is_same<_Ty, _Others>...);
-#endif // _LIBCUDACXX_STD_VER > 11
+#endif // _CCCL_STD_VER > 2011
 
 _LIBCUDACXX_END_NAMESPACE_STD