diff --git a/cmake.spec b/cmake.spec index 730398c3574..8226f074f2f 100644 --- a/cmake.spec +++ b/cmake.spec @@ -1,11 +1,14 @@ -### RPM external cmake 3.31.7 +### RPM external cmake 3.31.12 %define downloaddir %(echo %realversion | cut -d. -f1,2) Source: http://www.cmake.org/files/v%{downloaddir}/%n-%realversion.tar.gz +# Note: this patch should be removed after updating to CMake 4.4+ +Patch0: patches/cmake_cuda_std_23 Requires: bz2lib curl expat zlib BuildRequires: gmake %prep %setup -n cmake-%realversion +%patch0 -p1 %build cat > build-flags.cmake <<- EOF diff --git a/cuda-flags.file b/cuda-flags.file index 87cbcf721bc..ebe995cd488 100644 --- a/cuda-flags.file +++ b/cuda-flags.file @@ -2,8 +2,8 @@ ## INCLUDE cpp-standard # define the CUDA compilation flags in a way that can be shared by SCRAM-based and regular tools -# build support for Volta (7.0), Turing (7.5), Ampere (8.x), Lovelace (8.9), Hopper (9.0) and Blackwell (10.0, 12.0) -%define cuda_arch 70 75 80 89 90 100 120 +# build support for Turing (7.5), Ampere (8.x), Lovelace (8.9), Hopper (9.0) and Blackwell (10.0, 12.0) +%define cuda_arch 75 80 89 90 100 120 # LIBOMPTARGET_NVPTX_COMPUTE_CAPABILITIES style for listing the supported CUDA compute architectures %define omptarget_cuda_archs %(echo $(for ARCH in %cuda_arch; do echo "$ARCH"; done) | sed -e"s/ /,/g") diff --git a/cuda.spec b/cuda.spec index 5a55f848f1f..81b8ff09999 100644 --- a/cuda.spec +++ b/cuda.spec @@ -1,8 +1,8 @@ -### RPM external cuda 12.9.1 +### RPM external cuda 13.3.0 ## INITENV +PATH LD_LIBRARY_PATH %i/lib64 %define runpath_opts -m compute-sanitizer -m drivers -m nvvm -%define driversversion 575.57.08 +%define driversversion 610.43.02 %ifarch x86_64 Source0: https://developer.download.nvidia.com/compute/cuda/%{realversion}/local_installers/%{n}_%{realversion}_%{driversversion}_linux.run @@ -10,6 +10,7 @@ Source0: https://developer.download.nvidia.com/compute/cuda/%{realversion}/local %ifarch aarch64 Source0: https://developer.download.nvidia.com/compute/cuda/%{realversion}/local_installers/%{n}_%{realversion}_%{driversversion}_linux_sbsa.run %endif +Patch0: patches/cuda_cccl_8771 Requires: python3 AutoReq: no @@ -53,13 +54,11 @@ chmod a+x %_builddir/build/extras/CUPTI/lib64/*.so* mv %_builddir/build/extras/CUPTI/lib64/*.so* %{i}/lib64/ mv %_builddir/build/extras/CUPTI/include/*.h %{i}/include/ -# leave out the Nsight and NVVP graphical tools, and package the other binaries -rm -f %_builddir/build/bin/computeprof +# leave out the graphical tools, and package the other binaries rm -f %_builddir/build/bin/cuda-uninstaller rm -f %_builddir/build/bin/ncu* rm -f %_builddir/build/bin/nsight* rm -f %_builddir/build/bin/nsys* -rm -f %_builddir/build/bin/nvvp mv %_builddir/build/bin %{i}/ # package the cuda-gdb support files, and rename the binary to use it via a wrapper @@ -107,6 +106,10 @@ cp -p %_builddir/build/drivers/libnvidia-ml.so.%{driversversion} ln -sf libnvidia-ml.so.%{driversversion} %{i}/lib64/stubs/libnvidia-ml.so.1 ln -sf libnvidia-ml.so.1 %{i}/lib64/stubs/libnvidia-ml.so +# fix invalid C++ syntax in cub/device/device_transform.cuh +# backport NVIDIA/cccl#8771 from the main branch +patch -d %{i}/include/cccl -p2 < %{PATCH0} + %post # let nvcc find its components when invoked from the command line sed \ diff --git a/cudnn.spec b/cudnn.spec index b7cc5e08e71..a39bb084655 100644 --- a/cudnn.spec +++ b/cudnn.spec @@ -1,7 +1,7 @@ -### RPM external cudnn 9.9.0.52 +### RPM external cudnn 9.23.0.39 ## INITENV +PATH LD_LIBRARY_PATH %i/lib64 -%define cudaver 12 +%define cudaver 13 # NVIDIA uses sbsa for aarch64, and the standard architecture name for ppc64le and x86_64 %ifarch aarch64 diff --git a/gdrcopy.spec b/gdrcopy.spec index 8fa58f17a4e..f5dbdfa4438 100644 --- a/gdrcopy.spec +++ b/gdrcopy.spec @@ -1,4 +1,4 @@ -### RPM external gdrcopy 2.5.1 +### RPM external gdrcopy 2.6 ## INITENV +PATH LD_LIBRARY_PATH %i/lib64 Source: https://github.com/NVIDIA/%{n}/archive/v%{realversion}.tar.gz Requires: cuda diff --git a/onnxruntime/absl_nvidia_bug_6302392.patch b/onnxruntime/absl_nvidia_bug_6302392.patch new file mode 100644 index 00000000000..65f7262de2a --- /dev/null +++ b/onnxruntime/absl_nvidia_bug_6302392.patch @@ -0,0 +1,61 @@ +diff --git a/absl/container/internal/common.h b/absl/container/internal/common.h +index 5ef6c569..86de28c6 100644 +--- a/absl/container/internal/common.h ++++ b/absl/container/internal/common.h +@@ -68,6 +68,13 @@ struct IfRRef { + using AddPtr = Other*; + }; + ++// Workaround for NVCC (cudafe++) which fails to parse ++// `IfRRef::AddPtr` when used as a template argument ++// inside a heavily macro-expanded template parameter list. Top-level alias ++// defers the member-template lookup outside the surrounding template-id. ++template ++using IfRRefAddPtr = typename IfRRef::template AddPtr; ++ + template + struct IsTransparent : std::false_type {}; + template +diff --git a/absl/container/internal/raw_hash_map.h b/absl/container/internal/raw_hash_map.h +index b42a4f22..540788b7 100644 +--- a/absl/container/internal/raw_hash_map.h ++++ b/absl/container/internal/raw_hash_map.h +@@ -106,8 +106,8 @@ class raw_hash_map : public raw_hash_set { + typename K = key_type, class V = mapped_type, \ + ABSL_INTERNAL_IF_##KValue##_NOR_##VValue( \ + int = (EnableIf::AddPtr, \ +- IfRRef::AddPtr>>()), \ ++ IfRRefAddPtr, \ ++ IfRRefAddPtr>>()), \ + ABSL_INTERNAL_SINGLE_ARG( \ + int &..., \ + decltype(EnableIf>()) = \ +diff --git a/absl/container/internal/btree_container.h b/absl/container/internal/btree_container.h +index 21f00ae4..0a017f9f 100644 +--- a/absl/container/internal/btree_container.h ++++ b/absl/container/internal/btree_container.h +@@ -497,8 +497,8 @@ class btree_map_container : public btree_set_container { + typename K = key_type, class M, \ + ABSL_INTERNAL_IF_##KValue##_NOR_##MValue( \ + int = (EnableIf::AddPtr, \ +- IfRRef::AddPtr>>()), \ ++ IfRRefAddPtr, \ ++ IfRRefAddPtr>>()), \ + ABSL_INTERNAL_SINGLE_ARG( \ + int &..., \ + decltype(EnableIf>()) = \ +@@ -598,10 +598,10 @@ class btree_map_container : public btree_set_container { + ABSL_INTERNAL_IF_##KValue( \ + class... Args, \ + int = (EnableIf< \ +- LifetimeBoundK::AddPtr>>())), \ ++ LifetimeBoundK>>())), \ + ABSL_INTERNAL_IF_##KValue( \ + decltype(EnableIf::AddPtr>>()) = 0, \ ++ K, KValue, IfRRefAddPtr>>()) = 0, \ + class... Args), \ + std::enable_if_t::value, int> = \ + 0> \ diff --git a/onnxruntime/backport28736.patch b/onnxruntime/backport28736.patch new file mode 100644 index 00000000000..4a0ef79979f --- /dev/null +++ b/onnxruntime/backport28736.patch @@ -0,0 +1,29 @@ +diff --git a/onnxruntime/contrib_ops/cuda/quantization/matmul_4bits.cu b/onnxruntime/contrib_ops/cuda/quantization/matmul_4bits.cu +index cb9eab579a..ac93777a58 100644 +--- a/onnxruntime/contrib_ops/cuda/quantization/matmul_4bits.cu ++++ b/onnxruntime/contrib_ops/cuda/quantization/matmul_4bits.cu +@@ -281,8 +281,9 @@ __global__ void __launch_bounds__(kWarpSize* kColsPerThreadBlock) MatMulFloatInt + #define UnRollReduction(unroll_size) \ + do { \ + constexpr int kUnroll = unroll_size; \ +- constexpr int kUnrollMask = 0xffffffff & (~(kUnroll * k_per_iter - 1)); \ +- for (; k_id < (k & kUnrollMask); k_id += kUnroll * k_per_iter) { \ ++ constexpr int kUnrollStep = kUnroll * k_per_iter; \ ++ const int k_unroll_bound = k - k % kUnrollStep; \ ++ for (; k_id < k_unroll_bound; k_id += kUnrollStep) { \ + _Pragma("unroll") for (int i = 0; i < kUnroll; i++) { \ + uint32_t value = *(reinterpret_cast(b_data_quant + k_per_iter / 2 * i)); \ + T scale = b_scale_vec[t_meta_k + k_per_iter / block_size * i]; \ +diff --git a/onnxruntime/core/providers/cuda/tensor/concat_impl.cu b/onnxruntime/core/providers/cuda/tensor/concat_impl.cu +index f369ead0f9..8b27b0dd93 100644 +--- a/onnxruntime/core/providers/cuda/tensor/concat_impl.cu ++++ b/onnxruntime/core/providers/cuda/tensor/concat_impl.cu +@@ -80,7 +80,7 @@ Status ConcatSameConcatDimImpl(cudaStream_t stream, const size_t element_bytes, + template Status ConcatSameConcatDimImpl(cudaStream_t stream, const size_t element_bytes, + const int block_size_including_axis_dim, + const int block_size_inside_axis_dim, const int64_t concat_size, +- void* output_data, const void** input_data, ++ void* output_data, const void** const input_data, + const size_t output_size); + + // input tensor addresses passed by value diff --git a/onnxruntime/ciso646.patch b/onnxruntime/ciso646.patch new file mode 100644 index 00000000000..f029f5a485a --- /dev/null +++ b/onnxruntime/ciso646.patch @@ -0,0 +1,13 @@ +diff --git a/onnxruntime/core/providers/cuda/cuda_common.h b/onnxruntime/core/providers/cuda/cuda_common.h +index f90eb2813a..883491791f 100644 +--- a/onnxruntime/core/providers/cuda/cuda_common.h ++++ b/onnxruntime/core/providers/cuda/cuda_common.h +@@ -7,7 +7,7 @@ + + // The following three lines were copied from ABSL + // cutlass needs them, because cutlass uses "and"/"or" keywords +-#ifdef __cplusplus ++#if defined(__cplusplus) && __cplusplus < 202002L + #include + #endif + diff --git a/onnxruntime/spec b/onnxruntime/spec index 3e91e031f3d..b368db5a93c 100644 --- a/onnxruntime/spec +++ b/onnxruntime/spec @@ -6,6 +6,10 @@ %define tag v%{realversion} Source: git+https://github.com/%{github_user}/%{n}.git?obj=%{branch}/%{tag}&export=%{n}-%{realversion}&submodules=1&output=/%{n}-%{realversion}.tgz Patch0: onnxruntime/cms-changes +# Note: these two patches should likely be removed for onnxruntime 1.27+ +Patch1: onnxruntime/backport28736 +Patch2: onnxruntime/absl_nvidia_bug_6302392 +Patch3: onnxruntime/ciso646 BuildRequires: cmake ninja Requires: protobuf py3-numpy py3-onnx zlib libpng py3-pybind11 re2 eigen @@ -14,6 +18,8 @@ Requires: protobuf py3-numpy py3-onnx zlib libpng py3-pybind11 re2 eigen %prep %setup -q -n %{n}-%{realversion} %patch0 -p1 +%patch1 -p1 +%patch3 -p1 %build rm -rf ../build; mkdir ../build; cd ../build @@ -59,6 +65,9 @@ cmake ../%{n}-%{realversion}/cmake \ # -Wno-error=overloaded-virtual # See https://github.com/intel/neural-speed/issues/188 +# patch Abseil bundled with ONNX to workaround NVIDIA bug #6302392 +patch -d %{_builddir}/build/_deps/abseil_cpp-src -p1 < %{PATCH2} + ninja -v %{makeprocesses} python3 ../%{n}-%{realversion}/setup.py build diff --git a/patches/cmake_cuda_std_23.patch b/patches/cmake_cuda_std_23.patch new file mode 100644 index 00000000000..b80f7d0d921 --- /dev/null +++ b/patches/cmake_cuda_std_23.patch @@ -0,0 +1,40 @@ +commit d01c9a429b25fcea8f5b77547adc23f3fede0471 +Author: Arha Gatram +Date: Wed Jun 10 15:53:55 2026 +0000 + + CUDA: Add support for cuda_std_23 for nvcc 13.3+ + + CUDA 13.3 has been released with support for `-std=c++23`. + +diff --git a/Modules/Compiler/NVIDIA.cmake b/Modules/Compiler/NVIDIA.cmake +index 0d0f03ad00..5bf7044159 100644 +--- a/Modules/Compiler/NVIDIA.cmake ++++ b/Modules/Compiler/NVIDIA.cmake +@@ -46,6 +46,14 @@ macro(__compiler_nvidia_cxx_standards lang) + set(CMAKE_${lang}_STANDARD_LATEST 20) + endif() + endif() ++ ++ if (NOT CMAKE_${lang}_COMPILER_VERSION VERSION_LESS 13.3) ++ if(CMAKE_${lang}_SIMULATE_VERSION VERSION_GREATER_EQUAL 19.29.30129) ++ set(CMAKE_${lang}23_STANDARD_COMPILE_OPTION "-std=c++23") ++ set(CMAKE_${lang}23_EXTENSION_COMPILE_OPTION "-std=c++23") ++ set(CMAKE_${lang}_STANDARD_LATEST 23) ++ endif() ++ endif() + else() + set(CMAKE_${lang}03_STANDARD_COMPILE_OPTION "") + set(CMAKE_${lang}03_EXTENSION_COMPILE_OPTION "") +@@ -75,6 +83,12 @@ macro(__compiler_nvidia_cxx_standards lang) + set(CMAKE_${lang}20_EXTENSION_COMPILE_OPTION "-std=c++20") + set(CMAKE_${lang}_STANDARD_LATEST 20) + endif() ++ ++ if (NOT CMAKE_${lang}_COMPILER_VERSION VERSION_LESS 13.3) ++ set(CMAKE_${lang}23_STANDARD_COMPILE_OPTION "-std=c++23") ++ set(CMAKE_${lang}23_EXTENSION_COMPILE_OPTION "-std=c++23") ++ set(CMAKE_${lang}_STANDARD_LATEST 23) ++ endif() + endif() + + __compiler_check_default_language_standard(${lang} 6.0 03) diff --git a/patches/cuda_cccl_8771.patch b/patches/cuda_cccl_8771.patch new file mode 100644 index 00000000000..86d697787fe --- /dev/null +++ b/patches/cuda_cccl_8771.patch @@ -0,0 +1,37 @@ +diff --git a/cub/cub/device/device_transform.cuh b/cub/cub/device/device_transform.cuh +index 1b21230619d..64d26c81be8 100644 +--- a/cub/cub/device/device_transform.cuh ++++ b/cub/cub/device/device_transform.cuh +@@ -42,9 +42,12 @@ struct __return_constant + } // namespace detail + CUB_NAMESPACE_END + ++namespace cuda ++{ + template +-struct ::cuda::proclaims_copyable_arguments> : ::cuda::std::true_type ++struct proclaims_copyable_arguments> : ::cuda::std::true_type + {}; ++} // namespace cuda + + CUB_NAMESPACE_BEGIN + //! DeviceTransform provides device-wide, parallel operations for transforming elements tuple-wise from multiple input +diff --git a/cub/cub/device/dispatch/tuning/tuning_transform.cuh b/cub/cub/device/dispatch/tuning/tuning_transform.cuh +index 499ac333072..97073b360fb 100644 +--- a/cub/cub/device/dispatch/tuning/tuning_transform.cuh ++++ b/cub/cub/device/dispatch/tuning/tuning_transform.cuh +@@ -42,10 +42,12 @@ struct always_true_predicate + } // namespace detail::transform + CUB_NAMESPACE_END + ++namespace cuda ++{ + template <> +-struct ::cuda::proclaims_copyable_arguments +- : ::cuda::std::true_type ++struct proclaims_copyable_arguments : ::cuda::std::true_type + {}; ++} // namespace cuda + + CUB_NAMESPACE_BEGIN + namespace detail::transform diff --git a/patches/pytorch-typename-decltype.patch b/patches/pytorch-typename-decltype.patch new file mode 100644 index 00000000000..16532e5357f --- /dev/null +++ b/patches/pytorch-typename-decltype.patch @@ -0,0 +1,38 @@ +commit c2671b853e1553b258a30c15d66f9647faae7aee +Author: Tom Fogal <60981+tfogal@users.noreply.github.com> +Date: Mon May 11 15:55:08 2026 +0000 + + Couple naming updates (#180996) + + The template keyword in ivalue_inl.h is technically required for a correct parse; better to fix it now, before compilers start getting strict about it. + + The 'List' commit is just a consistency with surrounding code change. + Pull Request resolved: https://github.com/pytorch/pytorch/pull/180996 + Approved by: https://github.com/eqy + +diff --git a/aten/src/ATen/core/List_inl.h b/aten/src/ATen/core/List_inl.h +index 96f78faea22..c32e77f3d8e 100644 +--- a/aten/src/ATen/core/List_inl.h ++++ b/aten/src/ATen/core/List_inl.h +@@ -198,7 +198,7 @@ typename List::internal_const_reference_type List::operator[](size_type po + template + typename List::internal_reference_type List::operator[](size_type pos) { + static_cast(impl_->list.at(pos)); // Throw the exception if it is out of range. +- return {impl_->list.begin() + static_castlist)::difference_type>(pos)}; ++ return {impl_->list.begin() + static_cast(pos)}; + } + + template +diff --git a/aten/src/ATen/core/ivalue_inl.h b/aten/src/ATen/core/ivalue_inl.h +index 1d17fe06906..0e09cd8bc08 100644 +--- a/aten/src/ATen/core/ivalue_inl.h ++++ b/aten/src/ATen/core/ivalue_inl.h +@@ -1941,7 +1941,7 @@ Tuple generic_to_tuple_impl( + const ivalue::TupleElements& t, + std::index_sequence /*unused*/) { + return std::make_tuple( +- t[INDEX].to>()...); ++ t[INDEX].template to>()...); + } + } // namespace detail + diff --git a/pip/torch.file b/pip/torch.file index 38fcc97f8cc..10d5217817c 100644 --- a/pip/torch.file +++ b/pip/torch.file @@ -7,6 +7,7 @@ Source97: pip/files/torch-build-env Source98: FindEigen3.cmake Source99: scram-tools.file/tools/eigen/env Patch0: patches/pytorch-system-fmt +Patch1: patches/pytorch-typename-decltype BuildRequires: cmake ninja python-python3 py3-packaging Requires: fftw3 eigen fxdiv numactl openmpi protobuf psimd python3 py3-PyYAML