[rocm-libraries] ROCm/rocm-libraries#1769 (commit c3544e4)

Fix rocThrust 3.0.0 base NVCC job failures

This PR contains fixes for libcudacxx/libhipcxx compatibility in
rocThrust, forming part of the base for aligning rocThrust with Thrust
v3.0.0.

It also bumps rocThrust CMake version to 3.16, which is actually needed
for the build.

Author: yungshengtu

CMakeLists.txt

@@ -2,7 +2,7 @@
 # Copyright 2019-2025 Advanced Micro Devices, Inc.
 # ########################################################################
 
-cmake_minimum_required(VERSION 3.10.2 FATAL_ERROR)
+cmake_minimum_required(VERSION 3.16 FATAL_ERROR)
 
 # --------------------------------------
 # Update these variables at release time

cmake/DownloadProject.CMakeLists.cmake.in

@@ -1,7 +1,7 @@
 # Distributed under the OSI-approved MIT License.  See accompanying
 # file LICENSE or https://github.com/Crascit/DownloadProject for details.
 
-cmake_minimum_required(VERSION 3.10.2 FATAL_ERROR)
+cmake_minimum_required(VERSION 3.16 FATAL_ERROR)
 
 project(${DL_ARGS_PROJ}-download NONE)
 

extra/CMakeLists.txt

@@ -2,7 +2,7 @@
 # Copyright 2019-2025 Advanced Micro Devices, Inc.
 # ########################################################################
 
-cmake_minimum_required(VERSION 3.10.2 FATAL_ERROR)
+cmake_minimum_required(VERSION 3.16 FATAL_ERROR)
 
 # This project includes tests that should be run after
 # rocThrust is installed from package or using `make install`

testing/cmake/check_namespace.cmake

@@ -5,7 +5,7 @@
 # manually with:
 # cmake -D "Thrust_SOURCE_DIR=<thrust project root>" -P check_namespace.cmake
 
-cmake_minimum_required(VERSION 3.15)
+cmake_minimum_required(VERSION 3.16)
 
 set(exclusions
   # This defines the macros and must have bare namespace declarations:

thrust/detail/alignment.h

@@ -33,14 +33,8 @@
 #include _THRUST_STD_INCLUDE(type_traits) // For `std::alignment_of`.
 
 #if _THRUST_HAS_DEVICE_SYSTEM_STD
-#  if THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_CUDA
-#    include <cuda/cmath>
-#  else
-#    include <hip/cmath>
-#  endif
-#endif
-
-#if !_THRUST_HAS_DEVICE_SYSTEM_STD
+#  include _THRUST_LIBCXX_INCLUDE(cmath)
+#else
 #  include <rocprim/detail/various.hpp>
 #endif
 
@@ -87,11 +81,7 @@ THRUST_HOST_DEVICE T aligned_reinterpret_cast(U u)
 THRUST_HOST_DEVICE inline _THRUST_STD::size_t aligned_storage_size(_THRUST_STD::size_t n, _THRUST_STD::size_t align)
 {
 #if _THRUST_HAS_DEVICE_SYSTEM_STD
-#  if THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_CUDA
-  return ::cuda::ceil_div(n, align) * align;
-#  else
-  return ::hip::ceil_div(n, align) * align;
-#  endif
+  return _THRUST_LIBCXX::ceil_div(n, align) * align;
 #else
   return ::rocprim::detail::ceiling_div(n, align) * align;
 #endif

thrust/detail/attributes.h

@@ -0,0 +1,40 @@
+//===----------------------------------------------------------------------===//
+//
+// 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 DETAIL_ATTRIBUTES_H
+#define DETAIL_ATTRIBUTES_H
+
+#include <thrust/detail/config.h>
+#if _THRUST_HAS_DEVICE_SYSTEM_STD
+
+// clang-format off
+#  include _THRUST_STD_INCLUDE(__cccl/attributes.h)
+// clang-format on
+
+#  define THRUST_DECLSPEC_EMPTY_BASES _CCCL_DECLSPEC_EMPTY_BASES
+
+#else // !_THRUST_HAS_DEVICE_SYSTEM_STD
+
+#  ifdef __has_declspec_attribute
+#    define THRUST_HAS_DECLSPEC_ATTRIBUTE(__x) __has_declspec_attribute(__x)
+#  else // ^^^ __has_declspec_attribute ^^^ / vvv !__has_declspec_attribute vvv
+#    define THRUST_HAS_DECLSPEC_ATTRIBUTE(__x) 0
+#  endif // !__has_declspec_attribute
+
+// MSVC needs extra help with empty base classes
+#  if THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC || THRUST_HAS_DECLSPEC_ATTRIBUTE(empty_bases)
+#    define THRUST_DECLSPEC_EMPTY_BASES __declspec(empty_bases)
+#  else // !THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC
+#    define THRUST_DECLSPEC_EMPTY_BASES
+#  endif // THRUST_HOST_COMPILER == THRUST_HOST_COMPILER_MSVC
+
+#endif // _THRUST_HAS_DEVICE_SYSTEM_STD
+
+#endif // DETAIL_ATTRIBUTES_H

thrust/detail/config/compiler.h

@@ -21,22 +21,15 @@
 
 #pragma once
 
-#if defined(__HIP__)
-// Macro enables Device Malloc
-#  ifndef __HIP_ENABLE_DEVICE_MALLOC__
-#    define __HIP_ENABLE_DEVICE_MALLOC__ 1
-#  endif
-#  include <hip/hip_runtime.h>
-#endif
+// Internal config header that is only included through thrust/detail/config/config.h
 
-// For _CCCL_IMPLICIT_SYSTEM_HEADER
-#if defined(__CUDACC__)
-#  include <cuda/__cccl_config>
-#elif defined(__has_include)
-#  if __has_include(<cuda/__cccl_config>)
-#    include <cuda/__cccl_config>
-#  endif
-#endif
+#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
 
 // enumerate host compilers we know about
 //! deprecated [Since 2.7]

thrust/detail/config/config.h

@@ -20,11 +20,21 @@
 
 #pragma once
 
-#if defined(__CUDACC__)
+#include <thrust/detail/config/libcxx.h> // IWYU pragma: export
+
 // For _CCCL_IMPLICIT_SYSTEM_HEADER
-#  include <cuda/__cccl_config> // IWYU pragma: export
+#if _THRUST_HAS_DEVICE_SYSTEM_STD
+#  include _THRUST_LIBCXX_INCLUDE(__cccl_config) // IWYU pragma: export
 #endif
 
+#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
+
 // NOTE: The order of these #includes matters.
 
 #include <thrust/detail/config/compiler.h> // IWYU pragma: export
@@ -39,7 +49,6 @@
 #include <thrust/detail/config/diagnostic.h> // IWYU pragma: export
 #include <thrust/detail/config/execution_space.h> // IWYU pragma: export
 #include <thrust/detail/config/global_workarounds.h> // IWYU pragma: export
-#include <thrust/detail/config/libcxx.h> // IWYU pragma: export
 #include <thrust/detail/config/namespace.h> // IWYU pragma: export
 #include <thrust/detail/config/rtti.h> // IWYU pragma: export
 #include <thrust/detail/config/visibility.h> // IWYU pragma: export

thrust/detail/config/cpp_compatibility.h

@@ -78,7 +78,7 @@
 #    define THRUST_INCLUDE_DEVICE_CODE 0
 #    define THRUST_INCLUDE_HOST_CODE   1
 #  endif
-#elif THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_CUDA
+#else
 // These definitions were intended for internal use only and are now obsolete.
 // If you relied on them, consider porting your code to use the functionality
 // in libcu++'s <nv/target> header.

thrust/detail/config/libcxx.h

@@ -22,7 +22,7 @@
 
 #pragma once
 
-#include <thrust/detail/config/config.h>
+#include <thrust/detail/config/preprocessor.h>
 
 // This is a utility file that helps managing which
 // 'std' implementation we're using. The provided
@@ -33,33 +33,36 @@
 //     #include _THRUST_STD_INCLUDE(optional)
 //     using optional_int = _THRUST_STD::optional<int>;
 
-// When targeting CUDA, we want to use 'libcudacxx'. This
-// should always be available, since we are also dependent
-// on CCCL/thrust, which is a sibling of CCCL/libcudacxx.
-#if THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_CUDA
+// If 'libcudacxx' is available
+// clang-format off
+#if THRUST_HAS_INCLUDE(<cuda/std/cassert>)
+// clang-format on
+#  define _THRUST_LIBCXX_INCLUDE(LIB)   <cuda/LIB>
 #  define _THRUST_STD_INCLUDE(LIB)      <cuda/std/LIB>
+#  define _THRUST_LIBCXX                ::cuda
 #  define _THRUST_STD                   _CUDA_VSTD
 #  define _THRUST_HAS_DEVICE_SYSTEM_STD 1
 #  define _THRUST_STD_NAMESPACE_BEGIN   _LIBCUDACXX_BEGIN_NAMESPACE_STD
 #  define _THRUST_STD_NAMESPACE_END     _LIBCUDACXX_END_NAMESPACE_STD
 
-// When targeting HIP, we want to use 'libhipcxx' if
-// available. We can do this by checking the existance
-// of a known include.
-#elif THRUST_DEVICE_SYSTEM == THRUST_DEVICE_SYSTEM_HIP
-#  if defined(__has_include) && __has_include(<hip/std/cassert>)
-#    define _THRUST_STD_INCLUDE(LIB)      <hip/std/LIB>
-#    define _THRUST_STD                   ::hip::std
-#    define _THRUST_HAS_DEVICE_SYSTEM_STD 1
-#    define _THRUST_STD_NAMESPACE_BEGIN   _LIBCUDACXX_BEGIN_NAMESPACE_STD
-#    define _THRUST_STD_NAMESPACE_END     _LIBCUDACXX_END_NAMESPACE_STD
-#  endif
+// If 'libhipcxx' is available
+// clang-format off
+#elif THRUST_HAS_INCLUDE(<hip/std/cassert>)
+// clang-format on
+#  define _THRUST_LIBCXX_INCLUDE(LIB)   <hip/LIB>
+#  define _THRUST_STD_INCLUDE(LIB)      <hip/std/LIB>
+#  define _THRUST_LIBCXX                ::hip
+#  define _THRUST_STD                   ::hip::std
+#  define _THRUST_HAS_DEVICE_SYSTEM_STD 1
+#  define _THRUST_STD_NAMESPACE_BEGIN   _LIBCUDACXX_BEGIN_NAMESPACE_STD
+#  define _THRUST_STD_NAMESPACE_END     _LIBCUDACXX_END_NAMESPACE_STD
 #endif
 
-// If not using GPU backend or GPU vendor's STD is not
-// found, use fallback.
+// If 'libcudacxx' or 'libhipcxx' is not found, use fallback.
 #ifndef _THRUST_HAS_DEVICE_SYSTEM_STD
-#  define _THRUST_STD_INCLUDE(LIB)      <LIB>
+#  define _THRUST_LIBCXX_INCLUDE(LIB)
+#  define _THRUST_STD_INCLUDE(LIB) <LIB>
+#  define _THRUST_LIBCXX
 #  define _THRUST_STD                   ::std
 #  define _THRUST_HAS_DEVICE_SYSTEM_STD 0
 #  define _THRUST_STD_NAMESPACE_BEGIN \