Skip to content
Merged
Show file tree
Hide file tree
Changes from 4 commits
Commits
File filter

Filter by extension

Filter by extension

Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
9 changes: 8 additions & 1 deletion cmake/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -129,6 +129,8 @@ cmake_dependent_option(onnxruntime_USE_FPA_INTB_GEMM "Build FpA IntB gemm cuda k
option(onnxruntime_USE_INT4_KV_CACHE "Build cuda kernels for int4 kv cache" OFF)
option(onnxruntime_USE_FP8_KV_CACHE "Build cuda kernels for fp8 kv cache" ON)
option(onnxruntime_QUICK_BUILD "Speed up build by skipping some kernels for faster development" OFF)
# Raises the minimum driver to the CUDA 12.4 level (Linux >= 550.54.14, Windows >= 551.61); always on for CUDA >= 13.0.
cmake_dependent_option(onnxruntime_CUDA_FATBIN_COMPRESS_SIZE "Compress CUDA fatbins with -compress-mode=size" OFF "onnxruntime_USE_CUDA" OFF)

option(onnxruntime_BUILD_FOR_NATIVE_MACHINE "Enable this option for turning on optimization specific to this machine" OFF)
option(onnxruntime_USE_AVX "Use AVX instructions" OFF)
Expand Down Expand Up @@ -1526,7 +1528,12 @@ if (onnxruntime_USE_CUDA)
message(FATAL_ERROR "onnxruntime_USE_FP4_QMOE requires CUDA Toolkit version 12.8 or newer")
endif()

if(CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL "13.0")
if(onnxruntime_CUDA_FATBIN_COMPRESS_SIZE AND CMAKE_CUDA_COMPILER_VERSION VERSION_LESS "12.8")
message(FATAL_ERROR "onnxruntime_CUDA_FATBIN_COMPRESS_SIZE requires CUDA Toolkit version 12.8 or newer")
endif()

if(CMAKE_CUDA_COMPILER_VERSION VERSION_GREATER_EQUAL "13.0" OR onnxruntime_CUDA_FATBIN_COMPRESS_SIZE)
message(STATUS "Compressing CUDA fatbins with -compress-mode=size (requires a CUDA 12.4 or newer driver)")
set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} -Xfatbin=-compress-all -compress-mode=size")
else()
set(CMAKE_CUDA_FLAGS "${CMAKE_CUDA_FLAGS} -Xfatbin=-compress-all")
Expand Down
2 changes: 1 addition & 1 deletion cmake/external/cuda_configuration.cmake
Original file line number Diff line number Diff line change
Expand Up @@ -203,7 +203,7 @@ macro(setup_cuda_architectures)
endforeach()

# Enable accelerated features (like WGMMA, TMA and setmaxnreg) for SM >= 90.
set(ARCHITECTURES_WITH_ACCEL "90" "100" "101" "110" "120")
set(ARCHITECTURES_WITH_ACCEL "90" "100" "103" "110" "120" "121")
unset(CMAKE_CUDA_ARCHITECTURES_NORMALIZED)
foreach(CUDA_ARCH IN LISTS CMAKE_CUDA_ARCHITECTURES)
if(CUDA_ARCH MATCHES "^([0-9]+)f$")
Expand Down
11 changes: 11 additions & 0 deletions cmake/onnxruntime_providers_cuda_plugin.cmake
Original file line number Diff line number Diff line change
Expand Up @@ -123,6 +123,17 @@ onnxruntime_add_shared_library_module(onnxruntime_providers_cuda_plugin
${CUDA_PLUGIN_EP_CU_SRCS}
)

if(WIN32)
# Add version information to the packaged plugin DLL.
target_sources(onnxruntime_providers_cuda_plugin PRIVATE
"${ONNXRUNTIME_ROOT}/core/providers/cuda/onnxruntime_providers_cuda.rc")
target_compile_definitions(onnxruntime_providers_cuda_plugin PRIVATE
FILE_NAME=\"onnxruntime_providers_cuda.dll\")
elseif(UNIX AND NOT APPLE)
# The build output is packaged directly, so do not embed the build machine's CUDA path.
set_target_properties(onnxruntime_providers_cuda_plugin PROPERTIES SKIP_BUILD_RPATH TRUE)
endif()

# Mirror directory structure in the Visual Studio solution tree under "onnxruntime".
source_group(TREE ${ONNXRUNTIME_ROOT} PREFIX "onnxruntime" FILES ${CUDA_EP_CC_SRCS} ${CUDA_EP_CU_SRCS})
source_group(TREE ${ONNXRUNTIME_ROOT} PREFIX "onnxruntime" FILES ${CUDA_CONTRIB_OPS_CC_SRCS} ${CUDA_CONTRIB_OPS_CU_SRCS})
Expand Down
42 changes: 31 additions & 11 deletions tools/ci_build/github/azure-pipelines/plugin-cuda-pipeline.yml
Original file line number Diff line number Diff line change
Expand Up @@ -55,6 +55,22 @@ parameters:
- dev
default: dev

# CUDA 13.x always compresses fatbins with -compress-mode=size. Enabling this makes the
# CUDA 12.8 build match it, which shrinks the binary a lot but raises the minimum driver
# to the CUDA 12.4 level (Linux >= 550.54.14, Windows >= 551.61).
- name: enable_cuda_fatbin_size_compression
displayName: 'Compress CUDA fatbins for size (CUDA 12.8 only; needs r550+ driver)'
type: boolean
default: false

# The fpA_intB GEMM/GEMV kernels are the largest single component of the binary after
# flash attention. They are opt-in at run time (ORT_FPA_INTB_GEMM / ep.cuda.fpa_intb_gemm),
# but MatMulNBits still uses them automatically when the weights are prepacked.
- name: enable_fpa_intb_gemm
displayName: 'Build fpA_intB GEMM CUDA kernels'
type: boolean
default: true

- name: cmake_build_type
type: string
default: 'Release'
Expand All @@ -72,9 +88,9 @@ variables:
# Non-dev package versions (release, RC) must use Release build type
- name: invalidBuildTypeConfig
value: ${{ and(ne(parameters.package_type, 'dev'), ne(parameters.cmake_build_type, 'Release')) }}
# aarch64 is only available for CUDA 13.x
- name: invalidAArch64Config
value: ${{ and(eq(parameters.build_linux_aarch64, true), ne(parameters.cuda_version, '13.x')) }}
# ARM64 is only available for CUDA 13.x
- name: invalidArm64Config
value: ${{ and(or(eq(parameters.build_windows_arm64, true), eq(parameters.build_linux_aarch64, true)), ne(parameters.cuda_version, '13.x')) }}

extends:
template: v1/1ES.Official.PipelineTemplate.yml@1esPipelines
Expand Down Expand Up @@ -106,7 +122,7 @@ extends:

stages:
# Validate parameter combinations
- ${{ if or(eq(variables['invalidBuildTypeConfig'], 'True'), eq(variables['invalidAArch64Config'], 'True')) }}:
- ${{ if or(eq(variables['invalidBuildTypeConfig'], 'True'), eq(variables['invalidArm64Config'], 'True')) }}:
- stage: Validate_Parameters
displayName: 'Validate Parameters'
dependsOn: []
Expand All @@ -123,11 +139,11 @@ extends:
echo "##vso[task.logissue type=error]Non-dev package version requires Release build type."
exit 1
displayName: 'ERROR: Non-dev package version requires Release build type'
- ${{ if eq(variables['invalidAArch64Config'], 'True') }}:
- ${{ if eq(variables['invalidArm64Config'], 'True') }}:
- script: |
echo "##vso[task.logissue type=error]Linux aarch64 build is only available for CUDA 13.x."
echo "##vso[task.logissue type=error]Windows ARM64 and Linux aarch64 builds are only available for CUDA 13.x."
exit 1
displayName: 'ERROR: aarch64 requires CUDA 13.x'
displayName: 'ERROR: ARM64 requires CUDA 13.x'
- ${{ else }}:
- template: stages/plugin-cuda-packaging-stage.yml
parameters:
Expand All @@ -140,14 +156,18 @@ extends:
package_type: ${{ parameters.package_type }}
version_file: ${{ variables.epVersionFile }}
cmake_build_type: ${{ parameters.cmake_build_type }}
enable_cuda_fatbin_size_compression: ${{ parameters.enable_cuda_fatbin_size_compression }}
enable_fpa_intb_gemm: ${{ parameters.enable_fpa_intb_gemm }}
${{ if eq(parameters.cuda_version, '12.8') }}:
python_package_name: 'onnxruntime-ep-cuda12'
docker_base_image: 'onnxruntimebuildcache.azurecr.io/internal/azureml/onnxruntime/build/cuda12_x64_almalinux8_gcc14:20251017.1'
cmake_x64_cuda_archs: '61-real;75-real;86-real;89-real;120-real;120-virtual'
cmake_arm64_cuda_archs: '61-real;75-real;86-real;89-real;120-real;120-virtual'
cmake_windows_x64_cuda_archs: '61-real;75-real;86-real;89-real;120-real'
cmake_linux_x64_cuda_archs: '75-real;80-real;86-real;89-real;90-real;120-real'
${{ if eq(parameters.cuda_version, '13.x') }}:
python_package_name: 'onnxruntime-ep-cuda13'
docker_base_image: 'onnxruntimebuildcache.azurecr.io/internal/azureml/onnxruntime/build/cuda13_x64_almalinux8_gcc14:20251107.1'
docker_base_image_aarch64: 'onnxruntimebuildcache.azurecr.io/public/azureml/onnxruntime_build_cuda13_aarch64_almalinux9_gcc14:20260323.1'
cmake_x64_cuda_archs: '75-real;80-real;86-real;89-real;90-real;120-real;120-virtual'
cmake_arm64_cuda_archs: '110-real;120-real;121-real;120-virtual'
cmake_windows_x64_cuda_archs: '75-real;80-real;86-real;89-real;120-real'
cmake_windows_arm64_cuda_archs: '120-real;121-real'
cmake_linux_x64_cuda_archs: '75-real;80-real;86-real;89-real;90-real;120-real'
cmake_linux_aarch64_cuda_archs: '89-real;90-real;100-real;103-real;120-real;121-real'
Original file line number Diff line number Diff line change
Expand Up @@ -51,6 +51,16 @@ parameters:
- RelWithDebInfo
- MinSizeRel

- name: enable_cuda_fatbin_size_compression
type: boolean
displayName: 'Compress CUDA fatbins with -compress-mode=size'
default: false

- name: enable_fpa_intb_gemm
type: boolean
displayName: 'Build fpA_intB GEMM CUDA kernels'
default: true

- name: docker_base_image
type: string
displayName: 'Linux x86_64 docker base image'
Expand All @@ -61,15 +71,25 @@ parameters:
displayName: 'Linux aarch64 docker base image'
default: ''

- name: cmake_x64_cuda_archs
- name: cmake_windows_x64_cuda_archs
type: string
displayName: 'CMAKE_CUDA_ARCHITECTURES for Windows x64'
default: '61-real;75-real;86-real;89-real;120-real'

- name: cmake_windows_arm64_cuda_archs
type: string
displayName: 'CMAKE_CUDA_ARCHITECTURES for Windows ARM64'
default: '120-real;121-real'

- name: cmake_linux_x64_cuda_archs
type: string
displayName: 'CMAKE_CUDA_ARCHITECTURES for x64'
default: '61-real;75-real;86-real;89-real;120-real;120-virtual'
displayName: 'CMAKE_CUDA_ARCHITECTURES for Linux x64'
default: '75-real;80-real;86-real;89-real;90-real;120-real'

- name: cmake_arm64_cuda_archs
- name: cmake_linux_aarch64_cuda_archs
type: string
displayName: 'CMAKE_CUDA_ARCHITECTURES for ARM64'
default: '110-real;120-real;121-real;120-virtual'
displayName: 'CMAKE_CUDA_ARCHITECTURES for Linux aarch64'
default: '89-real;90-real;100-real;103-real;120-real;121-real'

- name: python_version
type: string
Expand Down Expand Up @@ -147,11 +167,13 @@ stages:
parameters:
arch: 'x64'
cuda_version: ${{ parameters.cuda_version }}
cmake_cuda_archs: ${{ parameters.cmake_x64_cuda_archs }}
cmake_cuda_archs: ${{ parameters.cmake_windows_x64_cuda_archs }}
package_version: ${{ parameters.package_type }}
version_file: ${{ parameters.version_file }}
python_package_name: ${{ parameters.python_package_name }}
cmake_build_type: ${{ parameters.cmake_build_type }}
enable_cuda_fatbin_size_compression: ${{ parameters.enable_cuda_fatbin_size_compression }}
enable_fpa_intb_gemm: ${{ parameters.enable_fpa_intb_gemm }}

# Windows ARM64
- ${{ if eq(parameters.build_windows_arm64, true) }}:
Expand All @@ -160,11 +182,13 @@ stages:
arch: 'arm64'
cuda_version: ${{ parameters.cuda_version }}
arm64_cuda_version: ${{ parameters.arm64_cuda_version }}
cmake_cuda_archs: ${{ parameters.cmake_arm64_cuda_archs }}
cmake_cuda_archs: ${{ parameters.cmake_windows_arm64_cuda_archs }}
package_version: ${{ parameters.package_type }}
version_file: ${{ parameters.version_file }}
python_package_name: ${{ parameters.python_package_name }}
cmake_build_type: ${{ parameters.cmake_build_type }}
enable_cuda_fatbin_size_compression: ${{ parameters.enable_cuda_fatbin_size_compression }}
enable_fpa_intb_gemm: ${{ parameters.enable_fpa_intb_gemm }}

# Linux x64
- ${{ if eq(parameters.build_linux_x64, true) }}:
Expand All @@ -174,7 +198,7 @@ stages:
arch: 'x64'
machine_pool: 'onnxruntime-Ubuntu2404-AMD-CPU'
cuda_version: ${{ parameters.cuda_version }}
cmake_cuda_archs: ${{ parameters.cmake_x64_cuda_archs }}
cmake_cuda_archs: ${{ parameters.cmake_linux_x64_cuda_archs }}
package_version: ${{ parameters.package_type }}
version_file: ${{ parameters.version_file }}
python_package_name: ${{ parameters.python_package_name }}
Expand All @@ -183,6 +207,8 @@ stages:
python_version: ${{ parameters.python_version }}
docker_python_exe_path: ${{ parameters.docker_python_exe_path }}
artifact_name: cuda_plugin_linux_x64
enable_cuda_fatbin_size_compression: ${{ parameters.enable_cuda_fatbin_size_compression }}
enable_fpa_intb_gemm: ${{ parameters.enable_fpa_intb_gemm }}

# Linux aarch64 (CUDA 13.x only)
- ${{ if eq(parameters.build_linux_aarch64, true) }}:
Expand All @@ -192,7 +218,7 @@ stages:
arch: 'aarch64'
machine_pool: 'onnxruntime-linux-ARM64-CPU-2019'
cuda_version: ${{ parameters.cuda_version }}
cmake_cuda_archs: ${{ parameters.cmake_arm64_cuda_archs }}
cmake_cuda_archs: ${{ parameters.cmake_linux_aarch64_cuda_archs }}
package_version: ${{ parameters.package_type }}
version_file: ${{ parameters.version_file }}
python_package_name: ${{ parameters.python_package_name }}
Expand All @@ -201,6 +227,8 @@ stages:
python_version: ${{ parameters.python_version }}
docker_python_exe_path: ${{ parameters.docker_python_exe_path }}
artifact_name: cuda_plugin_linux_aarch64
enable_cuda_fatbin_size_compression: ${{ parameters.enable_cuda_fatbin_size_compression }}
enable_fpa_intb_gemm: ${{ parameters.enable_fpa_intb_gemm }}

# NuGet packaging (runs after all platform builds)
- template: plugin-cuda-nuget-packaging-stage.yml
Expand Down
Original file line number Diff line number Diff line change
Expand Up @@ -57,6 +57,14 @@ parameters:
type: string
default: 'cuda_plugin_linux_x64'

- name: enable_cuda_fatbin_size_compression
type: boolean
default: false

- name: enable_fpa_intb_gemm
type: boolean
default: true

stages:
- stage: ${{ parameters.stage_name }}
dependsOn: []
Expand All @@ -77,6 +85,16 @@ stages:
artifactName: ${{ parameters.artifact_name }}
variables:
- template: ../templates/common-variables.yml
- name: FatbinCompressDefine
${{ if eq(parameters.enable_cuda_fatbin_size_compression, true) }}:
value: 'onnxruntime_CUDA_FATBIN_COMPRESS_SIZE=ON'
${{ else }}:
value: ''
- name: FpaIntBGemmDefine
${{ if eq(parameters.enable_fpa_intb_gemm, true) }}:
value: 'onnxruntime_USE_FPA_INTB_GEMM=ON'
${{ else }}:
value: 'onnxruntime_USE_FPA_INTB_GEMM=OFF'
steps:
- checkout: self
clean: true
Expand Down Expand Up @@ -115,7 +133,7 @@ stages:
workingDirectory: $(Build.SourcesDirectory)
displayName: 'Build CUDA Plugin (Python ${{ parameters.python_version }}, ${{ parameters.arch }}, CUDA ${{ parameters.cuda_version }})'
env:
EXTRA_CMAKE_DEFINES: $(PluginEpVersionDefine)
EXTRA_CMAKE_DEFINES: $(PluginEpVersionDefine) $(FatbinCompressDefine) $(FpaIntBGemmDefine)

- script: |
set -e -x
Expand All @@ -125,8 +143,19 @@ stages:
echo "Error: Expected plugin binary not found at '$plugin_path'. Failing build to avoid publishing an invalid package."
exit 1
fi

dynamic_search_path="$(readelf -d "$plugin_path" | sed -nE 's/.*\((RPATH|RUNPATH)\).*\[([^]]*)\].*/\2/p')"
if [[ -n "$dynamic_search_path" && ":$dynamic_search_path:" == *::* ]]; then
echo "Error: Plugin binary contains an empty RPATH/RUNPATH component: '$dynamic_search_path'"
exit 1
fi
if [[ "$dynamic_search_path" == *cuda* ]]; then
echo "Error: Plugin binary contains a hard-coded CUDA RPATH/RUNPATH: '$dynamic_search_path'"
exit 1
fi

cp "$plugin_path" "$(Build.ArtifactStagingDirectory)/bin/"
displayName: 'Copy plugin binaries'
displayName: 'Verify and copy plugin binaries'

- script: |
set -e -x
Expand Down
Loading
Loading