diff --git a/.github/workflows/android.yml b/.github/workflows/android.yml index 44fb2f29a7799..b144db7b92253 100644 --- a/.github/workflows/android.yml +++ b/.github/workflows/android.yml @@ -79,8 +79,8 @@ jobs: run: | set -e -x BINARY_SIZE_THRESHOLD_ARGS="" - echo "Binary size threshold in bytes: 1597440" - BINARY_SIZE_THRESHOLD_ARGS="--threshold_size_in_bytes 1597440" + echo "Binary size threshold in bytes: 1598464" + BINARY_SIZE_THRESHOLD_ARGS="--threshold_size_in_bytes 1598464" # Ensure ANDROID_NDK_HOME is available and get its real path if [ -z "$ANDROID_NDK_HOME" ]; then diff --git a/.github/workflows/linux_ci.yml b/.github/workflows/linux_ci.yml index d1e70354d4887..ec33cb1f5d3da 100644 --- a/.github/workflows/linux_ci.yml +++ b/.github/workflows/linux_ci.yml @@ -79,7 +79,7 @@ jobs: dockerfile_path: tools/ci_build/github/linux/docker/Dockerfile.manylinux2_28_cpu docker_image_repo: onnxruntimecpubuildpythonx64 # MATERIALIZE=ON needs no explicit pre-build pip install on this leg: the manylinux cp314 - # image already ships onnx==1.22.0 + numpy==2.4.2 (the exact gate pins) via its + # image already ships onnx==1.23.0 + numpy==2.4.2 (the exact gate pins) via its # requirements.txt (install_deps.sh installs it into cp311-cp314). Intentional, not an # omission -- unlike the QNN/android legs which run on bare agents and pip-install the pins. extra_build_flags: '--use_binskim_compliant_compile_flags --build_wheel --build_nuget --enable_transformers_tool_test --cmake_extra_defines onnxruntime_BUILD_BENCHMARKS=ON --cmake_extra_defines onnxruntime_MATERIALIZE_ONNX_NODE_TESTS=ON' diff --git a/cmake/deps.txt b/cmake/deps.txt index dd582d7651e16..3842e281473f5 100644 --- a/cmake/deps.txt +++ b/cmake/deps.txt @@ -34,10 +34,9 @@ microsoft_gsl;https://github.com/microsoft/GSL/archive/refs/tags/v4.2.1.zip;1094 microsoft_wil;https://github.com/microsoft/wil/archive/refs/tags/v1.0.250325.1.zip;826c8bd47c2258ec61b8b218e031e5b33d27f761 mimalloc;https://github.com/microsoft/mimalloc/archive/refs/tags/v2.1.1.zip;d5ee7d34223d0567892db5179849939c8769dc41 mp11;https://github.com/boostorg/mp11/archive/refs/tags/boost-1.82.0.zip;9bc9e01dffb64d9e0773b2e44d2f22c51aace063 -# ONNX pinned to the 1.22.0 final release (tag v1.22.0). rc2->final changed only CI -# workflow yml + VERSION_NUMBER (no operator-schema, opset, or testdata change), so it is -# pure version plumbing. The 7 CI requirements.txt files are pinned to onnx==1.22.0 (now on PyPI). -onnx;https://github.com/onnx/onnx/archive/refs/tags/v1.22.0.zip;2b2cd58ac7a26df5371266149e0c76776330cdf1 +# ONNX 1.23.0 final release (opset 28). +# The 7 CI requirements.txt files are pinned to onnx==1.23.0. +onnx;https://github.com/onnx/onnx/archive/refs/tags/v1.23.0.zip;75a298554b4fc8af73d35b657fad59c496844807 # Use the latest commit of 10.9-GA onnx_tensorrt;https://github.com/onnx/onnx-tensorrt/archive/d5dce67db7c2e64b07e055571f5ec06f7f254de2.zip;01114d3b67650857281fa50faa2e412130a63b69 protobuf;https://github.com/protocolbuffers/protobuf/releases/download/v33.6/protobuf-33.6.tar.gz;051fee744fcca04dab37eb009d49aa3e9546acda diff --git a/cmake/external/onnx b/cmake/external/onnx index 2bb50465112fe..ee3ccbd2b2344 160000 --- a/cmake/external/onnx +++ b/cmake/external/onnx @@ -1 +1 @@ -Subproject commit 2bb50465112feca9003e1ed654d77f01ff1415ca +Subproject commit ee3ccbd2b2344299d3a4506c2954a47b2181a485 diff --git a/cmake/onnxruntime_unittests.cmake b/cmake/onnxruntime_unittests.cmake index 1d01a4482e678..b38c21b39421d 100644 --- a/cmake/onnxruntime_unittests.cmake +++ b/cmake/onnxruntime_unittests.cmake @@ -1736,13 +1736,13 @@ if (NOT onnxruntime_ENABLE_TRAINING_TORCH_INTEROP) # coverage gap this feature exists to close. # --------------------------------------------------------------------------- if (onnxruntime_MATERIALIZE_ONNX_NODE_TESTS AND NOT CMAKE_SYSTEM_NAME STREQUAL "Emscripten") - # onnx version pin: derive from the archive URL in cmake/deps.txt (single source of truth). - string(REGEX MATCH "v([0-9]+\\.[0-9]+\\.[0-9]+)" _onnx_url_ver "${DEP_URL_onnx}") + # Derive the expected wheel version from the archive URL so this also works with installed ONNX packages. + string(REGEX MATCH "v([0-9]+\\.[0-9]+\\.[0-9]+)" _onnx_url_version "${DEP_URL_onnx}") if(CMAKE_MATCH_1) set(_onnx_pinned_version ${CMAKE_MATCH_1}) else() - message(FATAL_ERROR "Could not parse the pinned onnx version from DEP_URL_onnx='${DEP_URL_onnx}' " - "(expected a vX.Y.Z tag). Fix cmake/deps.txt or this regex.") + message(FATAL_ERROR "Could not parse the pinned ONNX version from DEP_URL_onnx='${DEP_URL_onnx}' " + "(expected a vX.Y.Z tag). Fix cmake/deps.txt or this parser.") endif() # Python interpreter is not guaranteed for static test-only builds (the top-level @@ -1770,27 +1770,10 @@ if (NOT onnxruntime_ENABLE_TRAINING_TORCH_INTEROP) " OR reconfigure with -Donnxruntime_MATERIALIZE_ONNX_NODE_TESTS=OFF (node-test coverage will be dropped).\n" " Details: ${_onnx_err}") endif() - # onnx version gate: HARD FAIL on a genuine mismatch, but RC / pre-release AWARE. - # ONNX's opset-bump workflow ships wheels like 1.23.0rc1 or 1.23.0.dev20240101 whose - # COMPILED opset registry already matches the formal 1.23.0 tag, so we compare on the - # RELEASE BASE (major.minor.micro) rather than the raw string. This mirrors - # materialize_onnx_node_tests.py::_release_base EXACTLY (regex ^(\d+)\.(\d+)\.(\d+), with a - # raw-string fallback when there is no leading X.Y.Z) so the cmake and Python layers agree: - # an rcN/.devN wheel of the pinned tag passes, while a real major/minor/micro mismatch - # (e.g. 1.21.x, or 1.23.0 when pinned at 1.22.0) still FATALs. Both sides are normalized; - # _onnx_pinned_version is already a clean X.Y.Z (parsed from the deps.txt vX.Y.Z tag), so - # normalizing it is a no-op kept only for symmetry with the Python two-sided compare. - string(REGEX MATCH "^[0-9]+\\.[0-9]+\\.[0-9]+" _onnx_ver_base "${_onnx_ver}") - if(_onnx_ver_base STREQUAL "") - set(_onnx_ver_base "${_onnx_ver}") - endif() - string(REGEX MATCH "^[0-9]+\\.[0-9]+\\.[0-9]+" _onnx_pin_base "${_onnx_pinned_version}") - if(_onnx_pin_base STREQUAL "") - set(_onnx_pin_base "${_onnx_pinned_version}") - endif() - if(NOT _onnx_ver_base STREQUAL _onnx_pin_base) + # The source and wheel are both final releases, so require exact version parity. + if(NOT _onnx_ver STREQUAL _onnx_pinned_version) message(FATAL_ERROR - "onnx ${_onnx_ver} (release base ${_onnx_ver_base}) != pinned ${_onnx_pinned_version} " + "onnx ${_onnx_ver} != pinned ${_onnx_pinned_version} " "(cmake/deps.txt). A mismatched wheel bakes the wrong opset/IR into the materialized " "corpus (silent drift).\n" " Fix: pip install onnx==${_onnx_pinned_version}") @@ -1868,6 +1851,11 @@ if (NOT onnxruntime_ENABLE_TRAINING_TORCH_INTEROP) VERBATIM) add_custom_target(onnx_node_tests_materialized ALL DEPENDS ${_materialized_node_root}/.stamp) + # Keep the corpus available for targeted onnx_test_runner builds as well as + # normal ALL builds. The runner is the direct C++ consumer of this artifact. + if(TARGET onnx_test_runner) + add_dependencies(onnx_test_runner onnx_node_tests_materialized) + endif() if (NOT onnxruntime_REDUCED_OPS_BUILD) # First-class ctest over the materialized node corpus (the durable replacement for the diff --git a/cmake/patches/onnx/onnx.patch b/cmake/patches/onnx/onnx.patch index 8b0068b7b0c80..e86a723d96a76 100644 --- a/cmake/patches/onnx/onnx.patch +++ b/cmake/patches/onnx/onnx.patch @@ -1,8 +1,8 @@ diff --git a/CMakeLists.txt b/CMakeLists.txt -index 791e24b..5f861a1 100644 +index 6f1d56949..7e6331dec 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt -@@ -46,6 +46,7 @@ option(ONNX_DISABLE_EXCEPTIONS "Disable exception handling." OFF) +@@ -53,6 +53,7 @@ option(ONNX_DISABLE_EXCEPTIONS "Disable exception handling." OFF) option(ONNX_DISABLE_STATIC_REGISTRATION "Disable static registration for ONNX operator schemas." OFF) option(ONNX_USE_UNITY_BUILD "Enable Unity (Jumbo) build for" OFF) option(ONNX_PYODIDE_BUILD "Build ONNX for Pyodide/Emscripten (wasm32)." OFF) @@ -10,7 +10,7 @@ index 791e24b..5f861a1 100644 option(ONNX_INSTALL "Install ONNX targets, headers, and CMake config files" ON) if(WIN32) option(ONNX_USE_MSVC_STATIC_RUNTIME "Build with MSVC static runtime. Ignored if CMAKE_MSVC_RUNTIME_LIBRARY is defined." OFF) -@@ -515,7 +516,22 @@ add_dependencies(onnx_proto onnx_proto_gen) +@@ -562,7 +563,23 @@ add_dependencies(onnx_proto onnx_proto_headers) # Core library. Hand-written sources come from target_sources(onnx ...) in the # subdirectories; the protobuf objects are pulled in from onnx_proto. add_library(onnx $) @@ -22,6 +22,7 @@ index 791e24b..5f861a1 100644 + # the full operator/schema sources via add_subdirectory(onnx). + target_sources(onnx PRIVATE + "${ONNX_ROOT}/onnx/defs/data_type_utils.cc" ++ "${ONNX_ROOT}/onnx/defs/doc_strings.cc" + ) + # Ensure ONNX_ML is treated as ON for minimal build consistency with ORT's file + set(ONNX_ML ON CACHE BOOL "Enable traditional ML API." FORCE) @@ -34,11 +35,37 @@ index 791e24b..5f861a1 100644 add_dependencies(onnx onnx_proto) set_target_properties(onnx PROPERTIES CXX_VISIBILITY_PRESET hidden) add_onnx_global_defines(onnx) +diff --git a/onnx/defs/doc_strings.cc b/onnx/defs/doc_strings.cc +index 976280392..8c22e55ff 100644 +--- a/onnx/defs/doc_strings.cc ++++ b/onnx/defs/doc_strings.cc +@@ -1705,6 +1705,8 @@ const char kDoc_DequantizeLinear_ver24[] = ""; + const char kDoc_RandomNormal_ver1[] = ""; + const char kDoc_Round_ver11[] = ""; + const char kDoc_SpaceToDepth_ver1[] = ""; ++const char kDoc_SpaceToDepth_ver28[] = ""; ++const char kDoc_DepthToSpace_ver28[] = ""; + const char kDoc_InstanceNormalization_ver6[] = ""; + const char kDoc_ThresholdedRelu_ver10[] = ""; + const char kDoc_Acosh_ver9[] = ""; +diff --git a/onnx/defs/nn/defs.cc b/onnx/defs/nn/defs.cc +index 4b58256e6..1a8103e7c 100644 +--- a/onnx/defs/nn/defs.cc ++++ b/onnx/defs/nn/defs.cc +@@ -2692,7 +2692,7 @@ ONNX_OPERATOR_SET_SCHEMA( + // positive value. + axis += input_ndim; + } +- if (axis < 0) { ++ if (axis < 0 || axis >= input_ndim) { + fail_shape_inference( + "Unexpected axis value (", + axis, diff --git a/onnx/defs/nn/old.cc b/onnx/defs/nn/old.cc -index 3887169..8432613 100644 +index cbb1532c0..6597a2ef9 100644 --- a/onnx/defs/nn/old.cc +++ b/onnx/defs/nn/old.cc -@@ -4078,7 +4078,6 @@ ONNX_OPERATOR_SET_SCHEMA( +@@ -4118,7 +4118,6 @@ ONNX_OPERATOR_SET_SCHEMA( GroupNormalization, 18, OpSchema() @@ -46,16 +73,40 @@ index 3887169..8432613 100644 .SetDoc(GroupNormalization_ver18_doc) .Attr("epsilon", "The epsilon value to use to avoid division by zero.", AttributeProto::FLOAT, 1e-5f) .Attr( -diff --git a/onnx/defs/nn/defs.cc b/onnx/defs/nn/defs.cc -index 0c271d5..6d64da7 100644 ---- a/onnx/defs/nn/defs.cc -+++ b/onnx/defs/nn/defs.cc -@@ -2670,7 +2670,7 @@ ONNX_OPERATOR_SET_SCHEMA( - // positive value. - axis += input_ndim; - } -- if (axis < 0) { -+ if (axis < 0 || axis >= input_ndim) { - fail_shape_inference( - "Unexpected axis value (", - axis, +diff --git a/onnx/defs/printer.cc b/onnx/defs/printer.cc +index 1d43bccb0..b6beed6f3 100644 +--- a/onnx/defs/printer.cc ++++ b/onnx/defs/printer.cc +@@ -9,6 +9,9 @@ + #include + #include + #include ++#include ++#include ++#include + #include + #include + #include +@@ -31,6 +34,14 @@ std::vector ParseRawIntData(const TensorProto& tensor) { + // point, the shortest form that round-trips through the parser. + template + void PrintNumber(std::ostream& os, T value) { ++#if defined(__APPLE__) && defined(__ENVIRONMENT_IPHONE_OS_VERSION_MIN_REQUIRED__) && \ ++ __ENVIRONMENT_IPHONE_OS_VERSION_MIN_REQUIRED__ < 160300 ++ // std::to_chars is unavailable when targeting iOS releases before 16.3. ++ std::ostringstream value_stream; ++ value_stream.imbue(std::locale::classic()); ++ value_stream << std::setprecision(std::numeric_limits::max_digits10) << value; ++ os << value_stream.str(); ++#else + if constexpr (std::is_floating_point_v) { + // Normalize NaN: MSVC's to_chars emits payload forms like "nan(ind)" + // that the parser does not accept. +@@ -42,6 +53,7 @@ void PrintNumber(std::ostream& os, T value) { + std::array buf{}; + const auto res = std::to_chars(buf.data(), buf.data() + buf.size(), value); + os.write(buf.data(), res.ptr - buf.data()); ++#endif + } + + bool IsValidIdentifier(const std::string& str) { diff --git a/cmake/vcpkg-ports/onnx/binskim.patch b/cmake/vcpkg-ports/onnx/binskim.patch index 8b0068b7b0c80..e86a723d96a76 100644 --- a/cmake/vcpkg-ports/onnx/binskim.patch +++ b/cmake/vcpkg-ports/onnx/binskim.patch @@ -1,8 +1,8 @@ diff --git a/CMakeLists.txt b/CMakeLists.txt -index 791e24b..5f861a1 100644 +index 6f1d56949..7e6331dec 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt -@@ -46,6 +46,7 @@ option(ONNX_DISABLE_EXCEPTIONS "Disable exception handling." OFF) +@@ -53,6 +53,7 @@ option(ONNX_DISABLE_EXCEPTIONS "Disable exception handling." OFF) option(ONNX_DISABLE_STATIC_REGISTRATION "Disable static registration for ONNX operator schemas." OFF) option(ONNX_USE_UNITY_BUILD "Enable Unity (Jumbo) build for" OFF) option(ONNX_PYODIDE_BUILD "Build ONNX for Pyodide/Emscripten (wasm32)." OFF) @@ -10,7 +10,7 @@ index 791e24b..5f861a1 100644 option(ONNX_INSTALL "Install ONNX targets, headers, and CMake config files" ON) if(WIN32) option(ONNX_USE_MSVC_STATIC_RUNTIME "Build with MSVC static runtime. Ignored if CMAKE_MSVC_RUNTIME_LIBRARY is defined." OFF) -@@ -515,7 +516,22 @@ add_dependencies(onnx_proto onnx_proto_gen) +@@ -562,7 +563,23 @@ add_dependencies(onnx_proto onnx_proto_headers) # Core library. Hand-written sources come from target_sources(onnx ...) in the # subdirectories; the protobuf objects are pulled in from onnx_proto. add_library(onnx $) @@ -22,6 +22,7 @@ index 791e24b..5f861a1 100644 + # the full operator/schema sources via add_subdirectory(onnx). + target_sources(onnx PRIVATE + "${ONNX_ROOT}/onnx/defs/data_type_utils.cc" ++ "${ONNX_ROOT}/onnx/defs/doc_strings.cc" + ) + # Ensure ONNX_ML is treated as ON for minimal build consistency with ORT's file + set(ONNX_ML ON CACHE BOOL "Enable traditional ML API." FORCE) @@ -34,11 +35,37 @@ index 791e24b..5f861a1 100644 add_dependencies(onnx onnx_proto) set_target_properties(onnx PROPERTIES CXX_VISIBILITY_PRESET hidden) add_onnx_global_defines(onnx) +diff --git a/onnx/defs/doc_strings.cc b/onnx/defs/doc_strings.cc +index 976280392..8c22e55ff 100644 +--- a/onnx/defs/doc_strings.cc ++++ b/onnx/defs/doc_strings.cc +@@ -1705,6 +1705,8 @@ const char kDoc_DequantizeLinear_ver24[] = ""; + const char kDoc_RandomNormal_ver1[] = ""; + const char kDoc_Round_ver11[] = ""; + const char kDoc_SpaceToDepth_ver1[] = ""; ++const char kDoc_SpaceToDepth_ver28[] = ""; ++const char kDoc_DepthToSpace_ver28[] = ""; + const char kDoc_InstanceNormalization_ver6[] = ""; + const char kDoc_ThresholdedRelu_ver10[] = ""; + const char kDoc_Acosh_ver9[] = ""; +diff --git a/onnx/defs/nn/defs.cc b/onnx/defs/nn/defs.cc +index 4b58256e6..1a8103e7c 100644 +--- a/onnx/defs/nn/defs.cc ++++ b/onnx/defs/nn/defs.cc +@@ -2692,7 +2692,7 @@ ONNX_OPERATOR_SET_SCHEMA( + // positive value. + axis += input_ndim; + } +- if (axis < 0) { ++ if (axis < 0 || axis >= input_ndim) { + fail_shape_inference( + "Unexpected axis value (", + axis, diff --git a/onnx/defs/nn/old.cc b/onnx/defs/nn/old.cc -index 3887169..8432613 100644 +index cbb1532c0..6597a2ef9 100644 --- a/onnx/defs/nn/old.cc +++ b/onnx/defs/nn/old.cc -@@ -4078,7 +4078,6 @@ ONNX_OPERATOR_SET_SCHEMA( +@@ -4118,7 +4118,6 @@ ONNX_OPERATOR_SET_SCHEMA( GroupNormalization, 18, OpSchema() @@ -46,16 +73,40 @@ index 3887169..8432613 100644 .SetDoc(GroupNormalization_ver18_doc) .Attr("epsilon", "The epsilon value to use to avoid division by zero.", AttributeProto::FLOAT, 1e-5f) .Attr( -diff --git a/onnx/defs/nn/defs.cc b/onnx/defs/nn/defs.cc -index 0c271d5..6d64da7 100644 ---- a/onnx/defs/nn/defs.cc -+++ b/onnx/defs/nn/defs.cc -@@ -2670,7 +2670,7 @@ ONNX_OPERATOR_SET_SCHEMA( - // positive value. - axis += input_ndim; - } -- if (axis < 0) { -+ if (axis < 0 || axis >= input_ndim) { - fail_shape_inference( - "Unexpected axis value (", - axis, +diff --git a/onnx/defs/printer.cc b/onnx/defs/printer.cc +index 1d43bccb0..b6beed6f3 100644 +--- a/onnx/defs/printer.cc ++++ b/onnx/defs/printer.cc +@@ -9,6 +9,9 @@ + #include + #include + #include ++#include ++#include ++#include + #include + #include + #include +@@ -31,6 +34,14 @@ std::vector ParseRawIntData(const TensorProto& tensor) { + // point, the shortest form that round-trips through the parser. + template + void PrintNumber(std::ostream& os, T value) { ++#if defined(__APPLE__) && defined(__ENVIRONMENT_IPHONE_OS_VERSION_MIN_REQUIRED__) && \ ++ __ENVIRONMENT_IPHONE_OS_VERSION_MIN_REQUIRED__ < 160300 ++ // std::to_chars is unavailable when targeting iOS releases before 16.3. ++ std::ostringstream value_stream; ++ value_stream.imbue(std::locale::classic()); ++ value_stream << std::setprecision(std::numeric_limits::max_digits10) << value; ++ os << value_stream.str(); ++#else + if constexpr (std::is_floating_point_v) { + // Normalize NaN: MSVC's to_chars emits payload forms like "nan(ind)" + // that the parser does not accept. +@@ -42,6 +53,7 @@ void PrintNumber(std::ostream& os, T value) { + std::array buf{}; + const auto res = std::to_chars(buf.data(), buf.data() + buf.size(), value); + os.write(buf.data(), res.ptr - buf.data()); ++#endif + } + + bool IsValidIdentifier(const std::string& str) { diff --git a/cmake/vcpkg-ports/onnx/fix-dependency-protobuf.patch b/cmake/vcpkg-ports/onnx/fix-dependency-protobuf.patch index 4a438def6cf1b..d1d96f7a61e9a 100644 --- a/cmake/vcpkg-ports/onnx/fix-dependency-protobuf.patch +++ b/cmake/vcpkg-ports/onnx/fix-dependency-protobuf.patch @@ -2,14 +2,14 @@ diff --git a/CMakeLists.txt b/CMakeLists.txt index 4a7df7d71..257d77e61 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt -@@ -159,6 +159,7 @@ if(NOT DEFINED CMAKE_POSITION_INDEPENDENT_CODE) +@@ -198,6 +198,7 @@ if(NOT DEFINED CMAKE_POSITION_INDEPENDENT_CODE) set(CMAKE_POSITION_INDEPENDENT_CODE ON) endif() +find_package(protobuf CONFIG REQUIRED) list(APPEND CMAKE_MODULE_PATH ${ONNX_ROOT}/cmake/external) - if(NOT ONNX_BUILD_CUSTOM_PROTOBUF) - if((ONNX_USE_LITE_PROTO AND TARGET protobuf::libprotobuf-lite) OR ((NOT ONNX_USE_LITE_PROTO) AND TARGET protobuf::libprotobuf)) + if(ONNX_USE_ASAN OR ONNX_USE_TSAN OR ONNX_USE_TYPESAN) + find_package(Sanitizer REQUIRED) diff --git a/cmake/ONNXConfig.cmake.in b/cmake/ONNXConfig.cmake.in index aa1ef47d9..d450b51e5 100644 --- a/cmake/ONNXConfig.cmake.in diff --git a/cmake/vcpkg-ports/onnx/portfile.cmake b/cmake/vcpkg-ports/onnx/portfile.cmake index 7a266e1a07663..60fa59000bb69 100644 --- a/cmake/vcpkg-ports/onnx/portfile.cmake +++ b/cmake/vcpkg-ports/onnx/portfile.cmake @@ -3,8 +3,8 @@ vcpkg_check_linkage(ONLY_STATIC_LIBRARY) vcpkg_from_github( OUT_SOURCE_PATH SOURCE_PATH REPO onnx/onnx - REF "v1.22.0" - SHA512 13fafff073a8e0bcf67fd06195da979c3c6273b3f5fdd25f5e5c39ad6af20eefe107f13dc73d52b6021e986d87401e46121e00bba8fbd09f098334e34f78462f + REF "v1.23.0" + SHA512 df13189eaf9d746341419f8067ada4ff0a44b4f09d2ed58c993eed8e1ada6c76d4318cc2bc63feaeb6e606804d567758b6eee4f620e05e23f92055d8141cce90 PATCHES fix-cmakelists.patch fix-dependency-protobuf.patch diff --git a/cmake/vcpkg-ports/onnx/vcpkg.json b/cmake/vcpkg-ports/onnx/vcpkg.json index c399d4c1ab087..ab822a1b6c87f 100644 --- a/cmake/vcpkg-ports/onnx/vcpkg.json +++ b/cmake/vcpkg-ports/onnx/vcpkg.json @@ -1,6 +1,6 @@ { "name": "onnx", - "version-semver": "1.22.0", + "version-semver": "1.23.0", "port-version": 0, "description": "Open standard for machine learning interoperability", "homepage": "https://onnx.ai", diff --git a/csharp/test/Microsoft.ML.OnnxRuntime.Tests.NetCoreApp/InferenceTest.netcore.cs b/csharp/test/Microsoft.ML.OnnxRuntime.Tests.NetCoreApp/InferenceTest.netcore.cs index c0475bb6102c1..14d4fffd55f56 100644 --- a/csharp/test/Microsoft.ML.OnnxRuntime.Tests.NetCoreApp/InferenceTest.netcore.cs +++ b/csharp/test/Microsoft.ML.OnnxRuntime.Tests.NetCoreApp/InferenceTest.netcore.cs @@ -491,9 +491,6 @@ private static Dictionary GetSkippedModels(DirectoryInfo modelsD { "test_mul_uint8", "Could not find an implementation for Mul(14) node with name" }, - { "test_bitshift_right_uint16", "Could not find an implementation for BitShift(11) nodeMeta with name ''"}, - { "test_bitshift_left_uint16", "Could not find an implementation for BitShift(11)"}, - { "test_pow_types_float32_uint64", "Could not find an implementation for Pow(15) node with name ''"}, { "test_pow_types_float32_uint32", "Could not find an implementation for Pow(15) node with name ''"}, diff --git a/docs/OperatorKernels.md b/docs/OperatorKernels.md index dfb50404021aa..b71601d4f34a8 100644 --- a/docs/OperatorKernels.md +++ b/docs/OperatorKernels.md @@ -63,7 +63,8 @@ The **OpSet Version** column uses the following notation: |||[9, 13]|**T** = tensor(double), tensor(float)| |||[7, 8]|**T** = tensor(double), tensor(float)| |BitCast|*in* input:**T1**
*out* output:**T2**|26+|**T1** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)
**T2** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| -|BitShift|*in* X:**T**
*in* Y:**T**
*out* Z:**T**|11+|**T** = tensor(uint32), tensor(uint64), tensor(uint8)| +|BitShift|*in* X:**T**
*in* Y:**T**
*out* Z:**T**|28+|**T** = tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| +|||[11, 27]|**T** = tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |BitwiseAnd|*in* A:**T**
*in* B:**T**
*out* C:**T**|18+|**T** = tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |BitwiseNot|*in* X:**T**
*out* Y:**T**|18+|**T** = tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |BitwiseOr|*in* A:**T**
*in* B:**T**
*out* C:**T**|18+|**T** = tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| @@ -287,7 +288,8 @@ The **OpSet Version** column uses the following notation: |||12|**T** = tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(int8), tensor(uint32), tensor(uint64), tensor(uint8)| |||[8, 11]|**T** = tensor(double), tensor(float)| |||[6, 7]|**T** = tensor(float)| -|Mod|*in* A:**T**
*in* B:**T**
*out* C:**T**|13+|**T** = tensor(double), tensor(float), tensor(float16), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| +|Mod|*in* A:**T**
*in* B:**T**
*out* C:**T**|28+|**T** = tensor(double), tensor(float), tensor(float16), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| +|||[13, 27]|**T** = tensor(double), tensor(float), tensor(float16), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |||[10, 12]|**T** = tensor(double), tensor(float), tensor(float16), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |Mul|*in* A:**T**
*in* B:**T**
*out* C:**T**|14+|**T** = tensor(double), tensor(float), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |||13|**T** = tensor(double), tensor(float), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| @@ -860,7 +862,8 @@ The **OpSet Version** column uses the following notation: |Min|*in* data_0:**T**
*out* min:**T**|13+|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| |||12|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| |||[6, 11]|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16)| -|Mod|*in* A:**T**
*in* B:**T**
*out* C:**T**|13+|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| +|Mod|*in* A:**T**
*in* B:**T**
*out* C:**T**|28+|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| +|||[13, 27]|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| |||[10, 12]|**T** = tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| |Mul|*in* A:**T**
*in* B:**T**
*out* C:**T**|14+|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |||13|**T** = tensor(bfloat16), tensor(double), tensor(float), tensor(float16), tensor(int32), tensor(int64), tensor(uint32), tensor(uint64)| @@ -1225,7 +1228,7 @@ The **OpSet Version** column uses the following notation: |||14+|**T** = tensor(float), tensor(float16)| |||9+|**T** = tensor(float), tensor(float16)| |||7+|**T** = tensor(float), tensor(float16)| -|BitShift|*in* X:**T**
*in* Y:**T**
*out* Z:**T**|11+|**T** = tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| +|BitShift|*in* X:**T**
*in* Y:**T**
*out* Z:**T**|[11, 27]|**T** = tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |BitwiseAnd|*in* A:**T**
*in* B:**T**
*out* C:**T**|18+|**T** = tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |BitwiseNot|*in* X:**T**
*out* Y:**T**|18+|**T** = tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| |BitwiseOr|*in* A:**T**
*in* B:**T**
*out* C:**T**|18+|**T** = tensor(int16), tensor(int32), tensor(int64), tensor(int8), tensor(uint16), tensor(uint32), tensor(uint64), tensor(uint8)| diff --git a/docs/python/examples/plot_pipeline.py b/docs/python/examples/plot_pipeline.py index e0bae57b7e2c8..c73d3ba30c211 100644 --- a/docs/python/examples/plot_pipeline.py +++ b/docs/python/examples/plot_pipeline.py @@ -29,12 +29,10 @@ ################################# -# Draw a model with ONNX +# Draw a model with pydot # ++++++++++++++++++++++ -# We use `net_drawer.py `_ -# included in *onnx* package. -# We use *onnx* to load the model -# in a different way than before. +# We use *onnx* to load the model in a different way than before, then +# construct a pydot graph from its nodes and values. from onnx import ModelProto # noqa: E402 @@ -46,18 +44,29 @@ ################################### # We convert it into a graph. -from onnx.tools.net_drawer import GetOpNodeProducer, GetPydotGraph # noqa: E402 +import pydot # noqa: E402 + +pydot_graph = pydot.Dot(model.graph.name, rankdir="LR") +value_nodes = {} +for index, node in enumerate(model.graph.node): + node_name = node.name or f"{node.op_type}_{index}" + op_node = pydot.Node(node_name, label=node.op_type, shape="box") + pydot_graph.add_node(op_node) + for value_name in node.input: + value_node = value_nodes.setdefault(value_name, pydot.Node(value_name, shape="ellipse")) + pydot_graph.add_node(value_node) + pydot_graph.add_edge(pydot.Edge(value_node, op_node)) + for value_name in node.output: + value_node = pydot.Node(value_name, shape="ellipse") + value_nodes[value_name] = value_node + pydot_graph.add_node(value_node) + pydot_graph.add_edge(pydot.Edge(op_node, value_node)) -pydot_graph = GetPydotGraph( - model.graph, name=model.graph.name, rankdir="LR", node_producer=GetOpNodeProducer("docstring") -) pydot_graph.write_dot("graph.dot") ####################################### # Then into an image -import os # noqa: E402 - -os.system("dot -O -Tpng graph.dot") +pydot_graph.write_png("graph.dot.png") ################################ # Which we display... diff --git a/include/onnxruntime/core/graph/schema_registry.h b/include/onnxruntime/core/graph/schema_registry.h index ca51e3621b2c6..f60589a6d82e0 100644 --- a/include/onnxruntime/core/graph/schema_registry.h +++ b/include/onnxruntime/core/graph/schema_registry.h @@ -37,7 +37,7 @@ class IOnnxRuntimeOpSchemaCollection : public ONNX_NAMESPACE::ISchemaRegistry { using ISchemaRegistry::GetSchema; - const ONNX_NAMESPACE::OpSchema* GetSchema(const std::string& key, const int maxInclusiveVersion, + const ONNX_NAMESPACE::OpSchema* GetSchema(const std::string& key, int maxInclusiveVersion, const std::string& domain) const final { const ONNX_NAMESPACE::OpSchema* latest_schema = nullptr; int earliest_opset_where_unchanged = std::numeric_limits::max(); diff --git a/include/onnxruntime/core/session/onnxruntime_c_api.h b/include/onnxruntime/core/session/onnxruntime_c_api.h index 4dd086c62cf96..6f1a7cf1fd6aa 100644 --- a/include/onnxruntime/core/session/onnxruntime_c_api.h +++ b/include/onnxruntime/core/session/onnxruntime_c_api.h @@ -223,6 +223,11 @@ typedef enum ONNXTensorElementDataType { ONNX_TENSOR_ELEMENT_DATA_TYPE_INT2, // maps to 4 packed int2 values (size == 1 byte) // Float8E8M0 type introduced in ONNX 1.21. 8-bit float with 8 exponent bits, 0 mantissa bits, no sign bit. ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT8E8M0, // Non-IEEE floating-point format, all values are powers of two + // Float6 types introduced in ONNX 1.23. + // ORT tensor storage uses one byte per element. TensorProto typed serialization uses int32_data; + // TensorProto raw_data uses bit-packed 6-bit values. + ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E2M3, + ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E3M2, } ONNXTensorElementDataType; // Synced with onnx TypeProto oneof diff --git a/js/web/docs/webgl-operators.md b/js/web/docs/webgl-operators.md index 1a8a99b88b2e0..2fbd03963809b 100644 --- a/js/web/docs/webgl-operators.md +++ b/js/web/docs/webgl-operators.md @@ -31,7 +31,7 @@ See [Compatibility](../README.md#Compatibility) for a list of the supported plat | [BitwiseOr](https://github.com/onnx/onnx/blob/main/docs/Operators.md#BitwiseOr) | | | [BitwiseXor](https://github.com/onnx/onnx/blob/main/docs/Operators.md#BitwiseXor) | | | [BlackmanWindow](https://github.com/onnx/onnx/blob/main/docs/Operators.md#BlackmanWindow) | | -| [Cast](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Cast) | [6-8](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-6), [9-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-9), [13-18](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-13), [19-20](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-19), [21-22](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-21), [23](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-23), [24](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-24), [25+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-25) | +| [Cast](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Cast) | [6-8](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-6), [9-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-9), [13-18](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-13), [19-20](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-19), [21-22](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-21), [23](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-23), [24](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-24), [25-27](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-25), [28+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Cast-28) | | [CastLike](https://github.com/onnx/onnx/blob/main/docs/Operators.md#CastLike) | | | [CausalConvWithState](https://github.com/onnx/onnx/blob/main/docs/Operators.md#CausalConvWithState) | | | [Ceil](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Ceil) | [6-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Ceil-6), [13+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Ceil-13) | @@ -53,7 +53,7 @@ See [Compatibility](../README.md#Compatibility) for a list of the supported plat | [CumSum](https://github.com/onnx/onnx/blob/main/docs/Operators.md#CumSum) | | | [DFT](https://github.com/onnx/onnx/blob/main/docs/Operators.md#DFT) | | | [DeformConv](https://github.com/onnx/onnx/blob/main/docs/Operators.md#DeformConv) | | -| [DepthToSpace](https://github.com/onnx/onnx/blob/main/docs/Operators.md#DepthToSpace) | [1-10](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#DepthToSpace-1), [11-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#DepthToSpace-11), [13+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#DepthToSpace-13) | +| [DepthToSpace](https://github.com/onnx/onnx/blob/main/docs/Operators.md#DepthToSpace) | [1-10](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#DepthToSpace-1), [11-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#DepthToSpace-11), [13-27](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#DepthToSpace-13), [28+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#DepthToSpace-28) | | [DequantizeLinear](https://github.com/onnx/onnx/blob/main/docs/Operators.md#DequantizeLinear) | | | [Det](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Det) | | | [Div](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Div) | [7-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Div-7), [13](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Div-13), [14+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Div-14) | @@ -144,7 +144,7 @@ See [Compatibility](../README.md#Compatibility) for a list of the supported plat | [Reciprocal](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Reciprocal) | | | [ReduceL1](https://github.com/onnx/onnx/blob/main/docs/Operators.md#ReduceL1) | | | [ReduceL2](https://github.com/onnx/onnx/blob/main/docs/Operators.md#ReduceL2) | | -| [ReduceLogSum](https://github.com/onnx/onnx/blob/main/docs/Operators.md#ReduceLogSum) | [1-10](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-1), [11-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-11), [13-17](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-13), [18+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-18) | +| [ReduceLogSum](https://github.com/onnx/onnx/blob/main/docs/Operators.md#ReduceLogSum) | [1-10](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-1), [11-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-11), [13-17](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-13), [18-27](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-18), [28+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceLogSum-28) | | [ReduceLogSumExp](https://github.com/onnx/onnx/blob/main/docs/Operators.md#ReduceLogSumExp) | | | [ReduceMax](https://github.com/onnx/onnx/blob/main/docs/Operators.md#ReduceMax) | [1-10](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMax-1), [11](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMax-11), [12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMax-12), [13-17](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMax-13), [18-19](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMax-18), [20+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMax-20) | | [ReduceMean](https://github.com/onnx/onnx/blob/main/docs/Operators.md#ReduceMean) | [1-10](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMean-1), [11-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMean-11), [13-17](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMean-13), [18+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#ReduceMean-18) | @@ -195,6 +195,7 @@ See [Compatibility](../README.md#Compatibility) for a list of the supported plat | [StringSplit](https://github.com/onnx/onnx/blob/main/docs/Operators.md#StringSplit) | | | [Sub](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Sub) | [7-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Sub-7), [13](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Sub-13), [14+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Sub-14) | | [Sum](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Sum) | [6-7](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Sum-6), [8-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Sum-8), [13+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Sum-13) | +| [SwiGLU](https://github.com/onnx/onnx/blob/main/docs/Operators.md#SwiGLU) | | | [Swish](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Swish) | | | [Tan](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Tan) | [7-21](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Tan-7), [22+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Tan-22) | | [Tanh](https://github.com/onnx/onnx/blob/main/docs/Operators.md#Tanh) | [6-12](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Tanh-6), [13+](https://github.com/onnx/onnx/blob/main/docs/Changelog.md#Tanh-13) | diff --git a/onnxruntime/contrib_ops/cuda/moe/moe_quantization.cc b/onnxruntime/contrib_ops/cuda/moe/moe_quantization.cc index 23f8c69fed824..6cc44c9e8eb05 100644 --- a/onnxruntime/contrib_ops/cuda/moe/moe_quantization.cc +++ b/onnxruntime/contrib_ops/cuda/moe/moe_quantization.cc @@ -1836,9 +1836,8 @@ Status QMoE::ComputeInternal(OpKernelContext* context) const { workspace_size, total_scratch_bytes); } #if !defined(BUILD_CUDA_EP_AS_PLUGIN) && !defined(ORT_MINIMAL_BUILD) - if (routing_record != nullptr) { - ORT_RETURN_IF_ERROR(routing_record->CaptureTile( - expert_indices, expert_scales, 0, SafeInt(num_rows) * SafeInt(k_), true, stream)); + if (routing_snapshot_) { + ORT_RETURN_IF_ERROR(routing_snapshot_->Consume()); } #endif return Status::OK(); diff --git a/onnxruntime/core/framework/onnxruntime_map_type_info.h b/onnxruntime/core/framework/onnxruntime_map_type_info.h index 71c680a2b1531..4e51aa00d3b3e 100644 --- a/onnxruntime/core/framework/onnxruntime_map_type_info.h +++ b/onnxruntime/core/framework/onnxruntime_map_type_info.h @@ -26,7 +26,7 @@ constexpr ONNXTensorElementDataType ToONNXTensorElementDataType( return ONNX_TENSOR_ELEMENT_DATA_TYPE_UNDEFINED; } - // The enums have the same values through FLOAT4E2M1. Their final three entries + // The enums have the same values through FLOAT4E2M1. The following three entries // differ because ONNX inserted FLOAT8E8M0 before UINT2 and INT2. switch (data_type) { case ONNX_NAMESPACE::TensorProto_DataType_FLOAT8E8M0: @@ -42,7 +42,7 @@ constexpr ONNXTensorElementDataType ToONNXTensorElementDataType( consteval bool IsTensorProtoToOrtElementTypeMapBijective() { constexpr size_t ort_element_type_count = - static_cast(ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT8E8M0) + 1; + static_cast(ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E3M2) + 1; constexpr size_t onnx_element_type_count = ONNX_NAMESPACE::TensorProto_DataType_DataType_ARRAYSIZE; if constexpr (onnx_element_type_count != ort_element_type_count) { return false; diff --git a/onnxruntime/core/framework/tensor_type_and_shape.cc b/onnxruntime/core/framework/tensor_type_and_shape.cc index e91d0396968e3..8f9ec22e49a3e 100644 --- a/onnxruntime/core/framework/tensor_type_and_shape.cc +++ b/onnxruntime/core/framework/tensor_type_and_shape.cc @@ -187,6 +187,12 @@ constexpr ONNXTensorElementDataType TensorDataTypeToOnnxRuntimeTensorElementData case o::TensorProto_DataType_FLOAT8E8M0: type = ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT8E8M0; break; + case o::TensorProto_DataType_FLOAT6E2M3: + type = ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E2M3; + break; + case o::TensorProto_DataType_FLOAT6E3M2: + type = ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E3M2; + break; case o::TensorProto_DataType_FLOAT16: type = ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT16; break; diff --git a/onnxruntime/core/framework/tensorprotoutils.cc b/onnxruntime/core/framework/tensorprotoutils.cc index 6a411f0d92efc..89d20c3a836ba 100644 --- a/onnxruntime/core/framework/tensorprotoutils.cc +++ b/onnxruntime/core/framework/tensorprotoutils.cc @@ -2119,6 +2119,8 @@ ONNXTensorElementDataType CApiElementTypeFromProtoType(int type) { #if !defined(DISABLE_FLOAT4_TYPES) CASE_TYPE(FLOAT4E2M1) #endif + CASE_TYPE(FLOAT6E2M3) + CASE_TYPE(FLOAT6E3M2) default: return ONNX_TENSOR_ELEMENT_DATA_TYPE_UNDEFINED; diff --git a/onnxruntime/core/framework/tensorprotoutils.h b/onnxruntime/core/framework/tensorprotoutils.h index 384136951ec7d..b8728b2f6c9ef 100644 --- a/onnxruntime/core/framework/tensorprotoutils.h +++ b/onnxruntime/core/framework/tensorprotoutils.h @@ -58,6 +58,8 @@ inline constexpr std::array path1{ - {0, 0, "Cast", {9, 13, 19, 21, 23, 24, 25}, kOnnxDomain}, + {0, 0, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}, kOnnxDomain}, {0, 0, "Slice", {10, 11, 13}, kOnnxDomain}, // Last Slice {0, 0, "Slice", {10, 11, 13}, kOnnxDomain}, // Mask Slice {0, 1, "Unsqueeze", {1, 11, 13, 21, 23, 24, 25}, kOnnxDomain}, diff --git a/onnxruntime/core/optimizer/embed_layer_norm_fusion.cc b/onnxruntime/core/optimizer/embed_layer_norm_fusion.cc index 0101a4f8b9da0..045ac89970b31 100644 --- a/onnxruntime/core/optimizer/embed_layer_norm_fusion.cc +++ b/onnxruntime/core/optimizer/embed_layer_norm_fusion.cc @@ -248,8 +248,8 @@ static bool MatchPositionEmbeddingSubgraphsFromGather( std::vector parent_path_1{ {0, 1, "Expand", {8, 13}, kOnnxDomain}, {0, 0, "Unsqueeze", {1, 11, 13, 21, 23, 24, 25}, kOnnxDomain}, - {0, 0, "Cast", {9, 13, 19, 21, 23, 24, 25}, kOnnxDomain}, - {0, 0, "Squeeze", {1, 11, 13, 21, 23, 24, 25}, kOnnxDomain}, + {0, 0, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}, kOnnxDomain}, + {0, 0, "Squeeze", {1, 11, 13, 21, 23, 24, 25, 28}, kOnnxDomain}, {0, 0, "Transpose", {1, 13, 21, 23, 24, 25}, kOnnxDomain}, {0, 0, "NonZero", {9, 13}, kOnnxDomain}, {0, 0, "ConstantOfShape", {9, 20, 21, 23, 24, 25}, kOnnxDomain}, @@ -260,7 +260,7 @@ static bool MatchPositionEmbeddingSubgraphsFromGather( std::vector parent_path_2{ {0, 1, "Expand", {8, 13}, kOnnxDomain}, {0, 0, "Unsqueeze", {1, 11, 13, 21, 23, 24, 25}, kOnnxDomain}, - {0, 0, "Squeeze", {1, 11, 13, 21, 23, 24, 25}, kOnnxDomain}, + {0, 0, "Squeeze", {1, 11, 13, 21, 23, 24, 25, 28}, kOnnxDomain}, {0, 0, "Transpose", {1, 13, 21, 23, 24, 25}, kOnnxDomain}, {0, 0, "NonZero", {9, 13}, kOnnxDomain}, {0, 0, "ConstantOfShape", {9, 20, 21, 23, 24, 25}, kOnnxDomain}, @@ -272,15 +272,15 @@ static bool MatchPositionEmbeddingSubgraphsFromGather( std::vector parent_path_3{ {0, 1, "Expand", {8, 13}, kOnnxDomain}, {0, 0, "Unsqueeze", {1, 11, 13, 21, 23, 24, 25}, kOnnxDomain}, - {0, 0, "Range", {11, 27}, kOnnxDomain}, - {0, 1, "Cast", {9, 13, 19, 21, 23, 24, 25}, kOnnxDomain}, + {0, 0, "Range", {11, 27, 28}, kOnnxDomain}, + {0, 1, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}, kOnnxDomain}, {0, 0, "Gather", {1, 11, 13}, kOnnxDomain}, {0, 0, "Shape", {1, 13, 15, 19, 21, 23, 24, 25}, kOnnxDomain}}; // Path 4 pattern (Path 3 with no "Cast"): std::vector parent_path_4{ {0, 1, "Expand", {8, 13}, kOnnxDomain}, {0, 0, "Unsqueeze", {1, 11, 13, 21, 23, 24, 25}, kOnnxDomain}, - {0, 0, "Range", {11, 27}, kOnnxDomain}, + {0, 0, "Range", {11, 27, 28}, kOnnxDomain}, {0, 1, "Gather", {1, 11, 13}, kOnnxDomain}, {0, 0, "Shape", {1, 13, 15, 19, 21, 23, 24, 25}, kOnnxDomain}}; // Match one of the three path patterns. diff --git a/onnxruntime/core/optimizer/fast_gelu_fusion.cc b/onnxruntime/core/optimizer/fast_gelu_fusion.cc index 13efcaabd83df..4909ae93e753f 100644 --- a/onnxruntime/core/optimizer/fast_gelu_fusion.cc +++ b/onnxruntime/core/optimizer/fast_gelu_fusion.cc @@ -175,7 +175,7 @@ MatchResult FastGeluFusion::CheckSecondFormula(Graph& graph, Node& pow1_node, if (p_cast1_node != nullptr) { Node& cast1_node = *graph.GetNode(p_cast1_node->Index()); // this is fused Cast node, so expect 2 output edges - if (!(graph_utils::IsSupportedOptypeVersionAndDomain(cast1_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + if (!(graph_utils::IsSupportedOptypeVersionAndDomain(cast1_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && cast1_node.InputDefs().size() == 1 && CheckNode(graph, cast1_node, pow1_node.GetExecutionProviderType(), false)) || cast1_node.GetOutputEdgesCount() != 2) { @@ -291,7 +291,7 @@ Status FastGeluFusion::ApplyImpl(Graph& graph, bool& modified, int graph_level, if (p_cast3_node == nullptr) continue; Node& cast3_node = *graph.GetNode(p_cast3_node->Index()); - if (!(graph_utils::IsSupportedOptypeVersionAndDomain(cast3_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + if (!(graph_utils::IsSupportedOptypeVersionAndDomain(cast3_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && cast3_node.InputDefs().size() == 1 && CheckNode(graph, cast3_node, node.GetExecutionProviderType(), true))) { continue; diff --git a/onnxruntime/core/optimizer/isinf_reducesum_fusion.cc b/onnxruntime/core/optimizer/isinf_reducesum_fusion.cc index 3f963e86acb0d..2185261629e24 100644 --- a/onnxruntime/core/optimizer/isinf_reducesum_fusion.cc +++ b/onnxruntime/core/optimizer/isinf_reducesum_fusion.cc @@ -45,7 +45,7 @@ Status IsInfReduceSumFusion::ApplyImpl(Graph& graph, bool& modified, int graph_l // This Cast can be skipped as we are replacing the subgraph with IsAllFinite, which supports FP16 auto cast1_node_iter = isinf_node.InputNodesBegin(); if (cast1_node_iter != isinf_node.InputNodesEnd() && - graph_utils::IsSupportedOptypeVersionAndDomain(*cast1_node_iter, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + graph_utils::IsSupportedOptypeVersionAndDomain(*cast1_node_iter, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && cast1_node_iter->GetOutputEdgesCount() == 1) { // check input type of cast node Node& cast1_node = *graph.GetNode(cast1_node_iter->Index()); @@ -65,7 +65,7 @@ Status IsInfReduceSumFusion::ApplyImpl(Graph& graph, bool& modified, int graph_l } Node& cast2_node = *graph.GetNode(cast2_node_itr->Index()); - if (!graph_utils::IsSupportedOptypeVersionAndDomain(cast2_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) || + if (!graph_utils::IsSupportedOptypeVersionAndDomain(cast2_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) || cast2_node.GetOutputEdgesCount() != 1 || graph.NodeProducesGraphOutput(cast2_node)) { continue; diff --git a/onnxruntime/core/optimizer/layer_norm_fusion.cc b/onnxruntime/core/optimizer/layer_norm_fusion.cc index e4716b5a0d366..9978175cce3ac 100644 --- a/onnxruntime/core/optimizer/layer_norm_fusion.cc +++ b/onnxruntime/core/optimizer/layer_norm_fusion.cc @@ -145,7 +145,7 @@ static bool IsPowExponentTwo(const Graph& graph, const Node& pow_node) { const Node* exponent_input_node = graph_utils::GetInputNode(pow_node, 1); if (exponent_input_node == nullptr || - !graph_utils::IsSupportedOptypeVersionAndDomain(*exponent_input_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) || + !graph_utils::IsSupportedOptypeVersionAndDomain(*exponent_input_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) || exponent_input_node->InputDefs().empty() || exponent_input_node->InputDefs()[0] == nullptr) { return false; } @@ -334,7 +334,7 @@ Status LayerNormFusion::ApplyImpl(Graph& graph, bool& modified, int graph_level, if (p_reduce_mean_input_node) { Node& reduce_mean_input_node = *graph.GetNode(p_reduce_mean_input_node->Index()); // If input to the 1st ReduceMean is a Cast, and the Cast has same consumer count as subCnt + 1 - if (graph_utils::IsSupportedOptypeVersionAndDomain(reduce_mean_input_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + if (graph_utils::IsSupportedOptypeVersionAndDomain(reduce_mean_input_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && reduce_mean_input_node.GetExecutionProviderType() == reduce_mean_node.GetExecutionProviderType() && optimizer_utils::CheckOutputEdges(graph, reduce_mean_input_node, static_cast(subCnt) + 1)) { nodes_to_remove.insert(nodes_to_remove.begin(), reduce_mean_input_node); @@ -347,7 +347,7 @@ Status LayerNormFusion::ApplyImpl(Graph& graph, bool& modified, int graph_level, const Node* p_cast1 = nullptr; if (!p_sub_node_dup && sub_node.GetOutputEdgesCount() == 1) { Node& cast_node = *graph.GetNode(sub_node.OutputNodesBegin()->Index()); - if (graph_utils::IsSupportedOptypeVersionAndDomain(cast_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + if (graph_utils::IsSupportedOptypeVersionAndDomain(cast_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && cast_node.GetExecutionProviderType() == reduce_mean_node.GetExecutionProviderType() && optimizer_utils::CheckOutputEdges(graph, cast_node, 2u) && IsSupportedDataType(cast_node)) { p_cast1 = &cast_node; @@ -446,7 +446,7 @@ Status LayerNormFusion::ApplyImpl(Graph& graph, bool& modified, int graph_level, const Node* p_cast2 = graph_utils::FirstParentByType(pow_node, "Cast"); if (p_cast2 != nullptr && p_cast2 != p_cast1) { Node& cast_node = *graph.GetNode(p_cast2->Index()); - if (!graph_utils::IsSupportedOptypeVersionAndDomain(cast_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) || + if (!graph_utils::IsSupportedOptypeVersionAndDomain(cast_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) || cast_node.GetExecutionProviderType() != reduce_mean_node.GetExecutionProviderType() || !optimizer_utils::CheckOutputEdges(graph, cast_node, 1)) { continue; @@ -464,7 +464,7 @@ Status LayerNormFusion::ApplyImpl(Graph& graph, bool& modified, int graph_level, // can be removed. This is one possible place a Cast Op can exist, that is between Div and Mul nodes. // div --> mul or div --> cast --> mul Node* next_node = graph.GetNode(div_node.OutputNodesBegin()->Index()); - if (graph_utils::IsSupportedOptypeVersionAndDomain(*next_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + if (graph_utils::IsSupportedOptypeVersionAndDomain(*next_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && optimizer_utils::CheckOutputEdges(graph, *next_node, 1)) { nodes_to_remove.push_back(*next_node); next_node = graph.GetNode(next_node->OutputNodesBegin()->Index()); @@ -735,7 +735,7 @@ Status SimplifiedLayerNormFusion::ApplyImpl(Graph& graph, bool& modified, int gr if (is_gpu_ep && p_pow_input_node) { Node& pow_input_node = *graph.GetNode(p_pow_input_node->Index()); // If input to Pow is a Cast, and the Cast has 2 consumers only (Pow, Div) - if (graph_utils::IsSupportedOptypeVersionAndDomain(pow_input_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + if (graph_utils::IsSupportedOptypeVersionAndDomain(pow_input_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && pow_input_node.GetExecutionProviderType() == pow_node.GetExecutionProviderType() && optimizer_utils::CheckOutputEdges(graph, pow_input_node, 2)) { nodes_to_remove.insert(nodes_to_remove.begin(), pow_input_node); @@ -745,7 +745,7 @@ Status SimplifiedLayerNormFusion::ApplyImpl(Graph& graph, bool& modified, int gr // div --> mul or div --> cast --> mul Node* next_node = graph.GetNode(div_node.OutputNodesBegin()->Index()); - if (graph_utils::IsSupportedOptypeVersionAndDomain(*next_node, "Cast", {9, 13, 19, 21, 23, 24, 25}) && + if (graph_utils::IsSupportedOptypeVersionAndDomain(*next_node, "Cast", {9, 13, 19, 21, 23, 24, 25, 28}) && optimizer_utils::CheckOutputEdges(graph, *next_node, 1)) { if (!is_gpu_ep) continue; nodes_to_remove.push_back(*next_node); diff --git a/onnxruntime/core/optimizer/pad_fusion.cc b/onnxruntime/core/optimizer/pad_fusion.cc index 179a08e0c6f9e..9a54f06c57f09 100644 --- a/onnxruntime/core/optimizer/pad_fusion.cc +++ b/onnxruntime/core/optimizer/pad_fusion.cc @@ -132,7 +132,7 @@ bool PadFusion::SatisfyCondition(const Graph& graph, const Node& node, const log } const Node& child_node = *node.OutputNodesBegin(); - if (graph_utils::IsSupportedOptypeVersionAndDomain(child_node, "Cast", {1, 6, 9, 13, 19, 21, 23, 24, 25})) { + if (graph_utils::IsSupportedOptypeVersionAndDomain(child_node, "Cast", {1, 6, 9, 13, 19, 21, 23, 24, 25, 28})) { if (child_node.GetOutputEdgesCount() != 1) { return false; } diff --git a/onnxruntime/core/optimizer/transpose_optimization/optimizer_api.h b/onnxruntime/core/optimizer/transpose_optimization/optimizer_api.h index cd0094877657d..ea3af9abd3bba 100644 --- a/onnxruntime/core/optimizer/transpose_optimization/optimizer_api.h +++ b/onnxruntime/core/optimizer/transpose_optimization/optimizer_api.h @@ -478,7 +478,7 @@ class GraphRef { } // namespace api constexpr int64_t kMinSupportedOpset = 7; -constexpr int64_t kMaxSupportedOpset = 27; +constexpr int64_t kMaxSupportedOpset = 28; // enum of results that a CostCheckFn can return. enum class CostCheckResult { diff --git a/onnxruntime/core/providers/coreml/builders/impl/base_op_builder.h b/onnxruntime/core/providers/coreml/builders/impl/base_op_builder.h index 609a965fa5198..0e49f488f8684 100644 --- a/onnxruntime/core/providers/coreml/builders/impl/base_op_builder.h +++ b/onnxruntime/core/providers/coreml/builders/impl/base_op_builder.h @@ -46,7 +46,7 @@ class BaseOpBuilder : public IOpBuilder { const logging::Logger& logger) const; virtual int GetMinSupportedOpSet(const Node& /*node*/) const { return 1; } - virtual int GetMaxSupportedOpSet(const Node& /*node*/) const { return 27; } + virtual int GetMaxSupportedOpSet(const Node& /*node*/) const { return 28; } bool HasSupportedOpSet(const Node& node, const logging::Logger& logger) const; bool HasSupportedInputs(const Node& node, const OpBuilderInputParams& input_params, diff --git a/onnxruntime/core/providers/cpu/cpu_execution_provider.cc b/onnxruntime/core/providers/cpu/cpu_execution_provider.cc index b91565dbc7f70..7d13ce4e16583 100644 --- a/onnxruntime/core/providers/cpu/cpu_execution_provider.cc +++ b/onnxruntime/core/providers/cpu/cpu_execution_provider.cc @@ -596,9 +596,10 @@ class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOn class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 12, MLFloat16, Gemm); #endif class ONNX_OPERATOR_VERSIONED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 12, GatherElements); -class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, uint8_t, BitShift); -class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, uint32_t, BitShift); -class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, uint64_t, BitShift); +class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 27, uint8_t, BitShift); +class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 27, uint16_t, BitShift); +class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 27, uint32_t, BitShift); +class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 27, uint64_t, BitShift); class ONNX_OPERATOR_VERSIONED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 12, Pad); class ONNX_OPERATOR_VERSIONED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 11, GatherND); class ONNX_OPERATOR_VERSIONED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 11, 26, Range); @@ -834,7 +835,8 @@ class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, int16_t, Neg); class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, int32_t, Neg); class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, int64_t, Neg); -class ONNX_OPERATOR_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, Mod); +class ONNX_OPERATOR_VERSIONED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, 27, Mod); +class ONNX_OPERATOR_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, Mod); class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, float, Abs); class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, double, Abs); class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 13, int8_t, Abs); @@ -1550,6 +1552,16 @@ class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, // Opset 27 class ONNX_OPERATOR_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 27, Range); +// Opset 28 +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, int8_t, BitShift); +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, int16_t, BitShift); +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, int32_t, BitShift); +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, int64_t, BitShift); +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, uint8_t, BitShift); +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, uint16_t, BitShift); +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, uint32_t, BitShift); +class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCpuExecutionProvider, kOnnxDomain, 28, uint64_t, BitShift); + // !!PLEASE READ BELOW!! Following that, add new entries above this comment /* *** IMPORTANT! *** @@ -2312,12 +2324,14 @@ Status RegisterOnnxOperatorKernels(KernelRegistry& kernel_registry) { double, Gemm)>, BuildKernelCreateInfo, - BuildKernelCreateInfo, - BuildKernelCreateInfo, - BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, BuildKernelCreateInfo, BuildKernelCreateInfo, @@ -2761,7 +2775,8 @@ Status RegisterOnnxOperatorKernels(KernelRegistry& kernel_registry) { BuildKernelCreateInfo, BuildKernelCreateInfo, BuildKernelCreateInfo, - BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, BuildKernelCreateInfo, BuildKernelCreateInfo, BuildKernelCreateInfo, @@ -3745,6 +3760,24 @@ Status RegisterOnnxOperatorKernels(KernelRegistry& kernel_registry) { // opset 27 BuildKernelCreateInfo, + + // opset 28 + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, }; for (auto& function_table_entry : function_table) { KernelCreateInfo info = function_table_entry(); diff --git a/onnxruntime/core/providers/cpu/math/element_wise_ops.cc b/onnxruntime/core/providers/cpu/math/element_wise_ops.cc index 4a9c4646a2c34..f4d8b9b6211b7 100644 --- a/onnxruntime/core/providers/cpu/math/element_wise_ops.cc +++ b/onnxruntime/core/providers/cpu/math/element_wise_ops.cc @@ -12,6 +12,9 @@ #include "core/mlas/inc/mlas.h" #include +#include +#include +#include namespace onnxruntime { // Supported types for operators that have type reduction enabled @@ -489,10 +492,18 @@ REG_ELEMENTWISE_VERSIONED_TYPED_KERNEL(Mean, 8, 12, float, Mean_8); // Supposed to add BFloat16 but we are not supporting now, however, separate registration REG_ELEMENTWISE_TYPED_KERNEL(Mean, 13, float, Mean_8); -REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 11, uint8_t, BitShift); -// REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 11, uint16_t, BitShift); -REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 11, uint32_t, BitShift); -REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 11, uint64_t, BitShift); +REG_ELEMENTWISE_VERSIONED_TYPED_KERNEL(BitShift, 11, 27, uint8_t, BitShift); +REG_ELEMENTWISE_VERSIONED_TYPED_KERNEL(BitShift, 11, 27, uint16_t, BitShift); +REG_ELEMENTWISE_VERSIONED_TYPED_KERNEL(BitShift, 11, 27, uint32_t, BitShift); +REG_ELEMENTWISE_VERSIONED_TYPED_KERNEL(BitShift, 11, 27, uint64_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, int8_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, int16_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, int32_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, int64_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, uint8_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, uint16_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, uint32_t, BitShift); +REG_ELEMENTWISE_TYPED_KERNEL(BitShift, 28, uint64_t, BitShift); REG_ELEMENTWISE_TYPED_KERNEL(BitwiseAnd, 18, int8_t, BitwiseAnd); REG_ELEMENTWISE_TYPED_KERNEL(BitwiseAnd, 18, int16_t, BitwiseAnd); @@ -1334,17 +1345,54 @@ BitShift::BitShift(const OpKernelInfo& info) : OpKernel(info) { ORT_THROW("Invalid direction value of '", direction, "'. Valid values are 'LEFT' or 'RIGHT'."); } -// Shifting by >= the bit width of an unsigned type is undefined behavior in C++. -// On x86, 64-bit shifts mask the shift amount to 6 bits, so shift by 64 acts like shift by 0. -// Guard against this by returning 0 when the shift amount >= the bit width. +template +T FromUnsignedBits(std::make_unsigned_t bits) { + static_assert(std::is_integral_v); + T value; + std::memcpy(&value, &bits, sizeof(value)); + return value; +} + +template +inline bool IsInvalidShift(T shift) { + using UnsignedT = std::make_unsigned_t; + constexpr UnsignedT bit_width = std::numeric_limits::digits; + return (std::is_signed_v && shift < 0) || static_cast(shift) >= bit_width; +} + template inline T SafeShiftLeft(T value, T shift) { - return shift >= sizeof(T) * 8 ? T{0} : value << shift; + if (IsInvalidShift(shift)) { + return T{0}; + } + + using UnsignedT = std::make_unsigned_t; + const auto result = static_cast( + static_cast(value) << static_cast(shift)); + return FromUnsignedBits(result); } template inline T SafeShiftRight(T value, T shift) { - return shift >= sizeof(T) * 8 ? T{0} : value >> shift; + if (IsInvalidShift(shift)) { + if constexpr (std::is_signed_v) { + return value < 0 ? T{-1} : T{0}; + } else { + return T{0}; + } + } + + using UnsignedT = std::make_unsigned_t; + constexpr UnsignedT bit_width = std::numeric_limits::digits; + const auto shift_amount = static_cast(shift); + auto result = static_cast(static_cast(value) >> shift_amount); + if constexpr (std::is_signed_v) { + if (value < 0 && shift_amount != 0) { + result |= std::numeric_limits::max() << (bit_width - shift_amount); + } + } + + return FromUnsignedBits(result); } template @@ -2110,9 +2158,19 @@ ONNX_CPU_OPERATOR_VERSIONED_KERNEL( BuildKernelDefConstraintsFromTypeList()), Mod); -ONNX_CPU_OPERATOR_KERNEL( +ONNX_CPU_OPERATOR_VERSIONED_KERNEL( Mod, 13, + 27, + KernelDefBuilder() + .TypeConstraint( + "T", + BuildKernelDefConstraintsFromTypeList()), + Mod); + +ONNX_CPU_OPERATOR_KERNEL( + Mod, + 28, KernelDefBuilder() .TypeConstraint( "T", @@ -2160,11 +2218,20 @@ void BroadCastFMod(OpKernelContext* context) { template inline T Modulus(T x, T y) { - auto res = x % y; + T res; + if constexpr (std::is_floating_point_v) { + res = std::fmod(x, y); + if (res == T{0}) { + return std::copysign(T{0}, y); + } + } else { + res = x % y; + } + if ((res < 0 && y > 0) || (res > 0 && y < 0)) { res += y; } - return static_cast(res); + return res; } template @@ -2242,6 +2309,42 @@ void BroadCastMLFloat16FMod(OpKernelContext* context) { UntypedBroadcastTwo(*context, funcs); } +void BroadCastMLFloat16Mod(OpKernelContext* context) { + ProcessBroadcastSpanFuncs funcs{ + [](BroadcastHelper& per_iter_bh) { + const auto X = per_iter_bh.ScalarInput0(); + auto Y = per_iter_bh.SpanInput1(); + auto output = per_iter_bh.OutputSpan(); + + std::transform(Y.begin(), Y.end(), output.begin(), + [X_fl = math::halfToFloat(X.val)](const MLFloat16& y) { + return MLFloat16(Modulus(X_fl, y.ToFloat())); + }); + }, + [](BroadcastHelper& per_iter_bh) { + auto X = per_iter_bh.SpanInput0(); + const MLFloat16 Y = per_iter_bh.ScalarInput1(); + auto output = per_iter_bh.OutputSpan(); + + std::transform(X.begin(), X.end(), output.begin(), + [Y_fl = math::halfToFloat(Y.val)](const MLFloat16& x) { + return MLFloat16(Modulus(x.ToFloat(), Y_fl)); + }); + }, + [](BroadcastHelper& per_iter_bh) { + auto X = per_iter_bh.SpanInput0(); + auto Y = per_iter_bh.SpanInput1(); + auto output = per_iter_bh.OutputSpan(); + + std::transform(X.begin(), X.end(), Y.begin(), output.begin(), + [](const MLFloat16& x, const MLFloat16& y) { + return MLFloat16(Modulus(x.ToFloat(), y.ToFloat())); + }); + }}; + + UntypedBroadcastTwo(*context, funcs); +} + template struct CallModImpl; @@ -2276,8 +2379,11 @@ struct CallModImpl::value> template struct CallModImpl::value, void>::type> { void operator()(bool fmod, OpKernelContext* ctx) const { - ORT_ENFORCE(fmod, "fmod attribute must be true for floating point types"); - BroadCastFMod(ctx); + if (fmod) { + BroadCastFMod(ctx); + } else { + BroadCastMod(ctx); + } } }; @@ -2285,8 +2391,11 @@ struct CallModImpl::value, template <> struct CallModImpl { void operator()(bool fmod, OpKernelContext* ctx) const { - ORT_ENFORCE(fmod, "fmod attribute must be true for floating point types"); - BroadCastMLFloat16FMod(ctx); + if (fmod) { + BroadCastMLFloat16FMod(ctx); + } else { + BroadCastMLFloat16Mod(ctx); + } } }; diff --git a/onnxruntime/core/providers/cuda/cu_inc/common.cuh b/onnxruntime/core/providers/cuda/cu_inc/common.cuh index f53da982538bd..a46385c9d2314 100644 --- a/onnxruntime/core/providers/cuda/cu_inc/common.cuh +++ b/onnxruntime/core/providers/cuda/cu_inc/common.cuh @@ -509,6 +509,40 @@ __device__ __inline__ T _Mod(T a, T b) { return r; } +template <> +__device__ __inline__ float _Mod(float a, float b) { + float r = fmodf(a, b); + if (r == 0.0f) { + return copysignf(0.0f, b); + } + if ((r > 0.0f && b < 0.0f) || (r < 0.0f && b > 0.0f)) { + r += b; + } + return r; +} + +template <> +__device__ __inline__ double _Mod(double a, double b) { + double r = fmod(a, b); + if (r == 0.0) { + return copysign(0.0, b); + } + if ((r > 0.0 && b < 0.0) || (r < 0.0 && b > 0.0)) { + r += b; + } + return r; +} + +template <> +__device__ __inline__ half _Mod(half a, half b) { + return half(_Mod(static_cast(a), static_cast(b))); +} + +template <> +__device__ __inline__ BFloat16 _Mod(BFloat16 a, BFloat16 b) { + return BFloat16(_Mod(static_cast(a), static_cast(b))); +} + template __device__ __inline__ T _Fmod(T a, T b) { return a % b; diff --git a/onnxruntime/core/providers/cuda/cuda_execution_provider.cc b/onnxruntime/core/providers/cuda/cuda_execution_provider.cc index e6645a0d4cb77..80dbd322fa534 100755 --- a/onnxruntime/core/providers/cuda/cuda_execution_provider.cc +++ b/onnxruntime/core/providers/cuda/cuda_execution_provider.cc @@ -1354,7 +1354,8 @@ class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, BFloat16, Tanh); class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, BFloat16, Gemm); class ONNX_OPERATOR_TYPED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, BFloat16, ReduceSum); -class ONNX_OPERATOR_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, Mod); +class ONNX_OPERATOR_VERSIONED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, 27, Mod); +class ONNX_OPERATOR_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 28, Mod); class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, 18, int8_t, QuantizeLinear); class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, 18, uint8_t, QuantizeLinear); class ONNX_OPERATOR_VERSIONED_TYPED_KERNEL_CLASS_NAME(kCudaExecutionProvider, kOnnxDomain, 13, 18, int8_t, DequantizeLinear); @@ -2647,7 +2648,8 @@ static Status RegisterCudaKernels(KernelRegistry& kernel_registry) { BuildKernelCreateInfo, BuildKernelCreateInfo, BuildKernelCreateInfo, - BuildKernelCreateInfo, + BuildKernelCreateInfo, + BuildKernelCreateInfo, BuildKernelCreateInfo, BuildKernelCreateInfo, BuildKernelCreateInfo, diff --git a/onnxruntime/core/providers/cuda/math/binary_elementwise_ops.cc b/onnxruntime/core/providers/cuda/math/binary_elementwise_ops.cc index babbb4b3ba672..ce4d089fa7b67 100644 --- a/onnxruntime/core/providers/cuda/math/binary_elementwise_ops.cc +++ b/onnxruntime/core/providers/cuda/math/binary_elementwise_ops.cc @@ -467,7 +467,14 @@ ONNX_OPERATOR_VERSIONED_KERNEL_EX( BuildKernelDefConstraints()), Mod); -ONNX_OPERATOR_KERNEL_EX(Mod, kOnnxDomain, 13, kCudaExecutionProvider, +ONNX_OPERATOR_VERSIONED_KERNEL_EX( + Mod, kOnnxDomain, 13, 27, kCudaExecutionProvider, + (*KernelDefBuilder::Create()) + .TypeConstraint("T", BuildKernelDefConstraints()), + Mod); + +ONNX_OPERATOR_KERNEL_EX(Mod, kOnnxDomain, 28, kCudaExecutionProvider, (*KernelDefBuilder::Create()) .TypeConstraint("T", BuildKernelDefConstraints()), @@ -478,10 +485,6 @@ Status Mod::ComputeInternal(OpKernelContext* context) const { BinaryElementwisePreparation prepare; ORT_RETURN_IF_ERROR(Prepare(context, &prepare)); auto element_type = prepare.lhs_tensor->GetElementType(); - ORT_ENFORCE(fmod_ || element_type == on::TensorProto_DataType_INT32 || - element_type == on::TensorProto_DataType_INT64 || element_type == on::TensorProto_DataType_UINT32 || - element_type == on::TensorProto_DataType_UINT64, - "Non-fmod can support integer types only."); #define CASE_MOD_ELEMENT_TYPE(name, onnx_type, data_type) \ case onnx_type: { \ Impl_##name::MappedType>( \ @@ -514,6 +517,10 @@ Status Mod::ComputeInternal(OpKernelContext* context) const { CASE_MOD_ELEMENT_TYPE(Mod, on::TensorProto_DataType_INT64, int64_t); CASE_MOD_ELEMENT_TYPE(Mod, on::TensorProto_DataType_UINT32, uint32_t); CASE_MOD_ELEMENT_TYPE(Mod, on::TensorProto_DataType_UINT64, uint64_t); + CASE_MOD_ELEMENT_TYPE(Mod, on::TensorProto_DataType_FLOAT, float); + CASE_MOD_ELEMENT_TYPE(Mod, on::TensorProto_DataType_DOUBLE, double); + CASE_MOD_ELEMENT_TYPE(Mod, on::TensorProto_DataType_FLOAT16, MLFloat16); + CASE_MOD_ELEMENT_TYPE(Mod, on::TensorProto_DataType_BFLOAT16, BFloat16); default: return ORT_MAKE_STATUS(ONNXRUNTIME, INVALID_ARGUMENT, "Unsupported element type: ", DataTypeImpl::ToString(prepare.lhs_tensor->DataType())); diff --git a/onnxruntime/core/providers/cuda/math/binary_elementwise_ops_impl.cu b/onnxruntime/core/providers/cuda/math/binary_elementwise_ops_impl.cu index 4e71a9f7c5090..2b08614b1e9c4 100644 --- a/onnxruntime/core/providers/cuda/math/binary_elementwise_ops_impl.cu +++ b/onnxruntime/core/providers/cuda/math/binary_elementwise_ops_impl.cu @@ -162,7 +162,7 @@ SPECIALIZED_BINARY_ELEMENTWISE_IMPL(Xor, bool) SPECIALIZED_BINARY_ELEMENTWISE_IMPL_HFD(PRelu) SPECIALIZED_BINARY_ELEMENTWISE_IMPL_UZILHFD(Max) SPECIALIZED_BINARY_ELEMENTWISE_IMPL_UZILHFD(Min) -SPECIALIZED_BINARY_ELEMENTWISE_IMPL_UZIL(Mod) +SPECIALIZED_BINARY_ELEMENTWISE_IMPL_UZILHFD(Mod) SPECIALIZED_BINARY_ELEMENTWISE_IMPL_UZILHFD(Fmod) // create declarations for impl for Pow diff --git a/onnxruntime/core/providers/dml/DmlExecutionProvider/src/AbiCustomRegistry.cpp b/onnxruntime/core/providers/dml/DmlExecutionProvider/src/AbiCustomRegistry.cpp index 076027dd3672f..3edc8d216f7c1 100644 --- a/onnxruntime/core/providers/dml/DmlExecutionProvider/src/AbiCustomRegistry.cpp +++ b/onnxruntime/core/providers/dml/DmlExecutionProvider/src/AbiCustomRegistry.cpp @@ -391,6 +391,16 @@ HRESULT STDMETHODCALLTYPE AbiCustomRegistry::RegisterOperatorKernel( builder.SetDomain(opKernel->domain) .SinceVersion(opKernel->minimumOperatorSetVersion) .Provider(providerType); + if (isInternalOperator && + opKernel->executionType == MLOperatorExecutionType::D3D12 && + opKernel->domain != nullptr && + std::string_view(opKernel->domain).empty() && + std::string_view(opKernel->name) == "BitShift" && + opKernel->minimumOperatorSetVersion == 11) + { + // DirectML masks out-of-range shift counts, which does not satisfy the opset-28 semantics. + builder.SinceVersion(11, 27); + } std::string_view name(opKernel->name); if (name == "MemcpyToHost") diff --git a/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/base_op_builder.h b/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/base_op_builder.h index a573b73d2bbb5..c52458b7046c5 100644 --- a/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/base_op_builder.h +++ b/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/base_op_builder.h @@ -72,7 +72,7 @@ class BaseOpBuilder : public IOpBuilder { const OpSupportCheckParams& params) const; virtual int GetMinSupportedOpSet(const NodeUnit& /* node_unit */) const { return 1; } - virtual int GetMaxSupportedOpSet(const NodeUnit& /* node_unit */) const { return 27; } + virtual int GetMaxSupportedOpSet(const NodeUnit& /* node_unit */) const { return 28; } // Check if this node_unit's type is supported // SingleNode type NodeUnit is supported diff --git a/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/dequantizelinear_op_builder.cc b/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/dequantizelinear_op_builder.cc index ed9062f894660..72544d87413cd 100644 --- a/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/dequantizelinear_op_builder.cc +++ b/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/dequantizelinear_op_builder.cc @@ -37,6 +37,8 @@ class DequantizeLinearOpBuilder : public BaseOpBuilder { return ANEURALNETWORKS_FEATURE_LEVEL_1; } + int GetMaxSupportedOpSet(const NodeUnit& /* node_unit */) const override { return 27; } + bool HasSupportedInputOutputsImpl( const GraphViewer& graph_viewer, const NodeUnit& node_unit, const OpSupportCheckParams& params) const override { diff --git a/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/quantizelinear_op_builder.cc b/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/quantizelinear_op_builder.cc index eb81f5e3f59ee..9a94f37b68ae6 100644 --- a/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/quantizelinear_op_builder.cc +++ b/onnxruntime/core/providers/nnapi/nnapi_builtin/builders/impl/quantizelinear_op_builder.cc @@ -37,6 +37,8 @@ class QuantizeLinearOpBuilder : public BaseOpBuilder { return ANEURALNETWORKS_FEATURE_LEVEL_3; } + int GetMaxSupportedOpSet(const NodeUnit& /* node_unit */) const override { return 27; } + bool HasSupportedInputOutputsImpl( const GraphViewer& graph_viewer, const NodeUnit& node_unit, const OpSupportCheckParams& params) const override { diff --git a/onnxruntime/core/providers/vsinpu/builders/impl/base_op_builder.h b/onnxruntime/core/providers/vsinpu/builders/impl/base_op_builder.h index 488581dbd0e97..8a9f342871f3d 100644 --- a/onnxruntime/core/providers/vsinpu/builders/impl/base_op_builder.h +++ b/onnxruntime/core/providers/vsinpu/builders/impl/base_op_builder.h @@ -49,7 +49,7 @@ class BaseOpBuilder : public IOpBuilder { virtual bool IsQuantizedOp(const NodeUnit& /* node_unit */) const { return false; } virtual int GetMinSupportedOpSet(const NodeUnit& /* node_unit */) const { return 1; } - virtual int GetMaxSupportedOpSet(const NodeUnit& /* node_unit */) const { return 27; } + virtual int GetMaxSupportedOpSet(const NodeUnit& /* node_unit */) const { return 28; } virtual bool HasSupportedInputOutputsImpl( const InitializedTensorSet& initializers, const NodeUnit& node_unit) const; diff --git a/onnxruntime/core/providers/webgpu/compute_context.h b/onnxruntime/core/providers/webgpu/compute_context.h index e151258a63a5b..edac52acf75d6 100644 --- a/onnxruntime/core/providers/webgpu/compute_context.h +++ b/onnxruntime/core/providers/webgpu/compute_context.h @@ -126,15 +126,13 @@ class ComputeContextBase { // // Get the logger. // + inline const auto& Logger() const { #if defined(ORT_USE_EP_API_ADAPTERS) - inline const onnxruntime::ep::adapter::Logger& Logger() const { return ep_.GetEpLogger(); - } #else - inline const logging::Logger& Logger() const { return *ep_.GetLogger(); - } #endif + } // // Run a compute shader program. diff --git a/onnxruntime/core/providers/webnn/builders/impl/base_op_builder.h b/onnxruntime/core/providers/webnn/builders/impl/base_op_builder.h index f4429fbb06457..d1334ba9a5420 100644 --- a/onnxruntime/core/providers/webnn/builders/impl/base_op_builder.h +++ b/onnxruntime/core/providers/webnn/builders/impl/base_op_builder.h @@ -52,7 +52,7 @@ class BaseOpBuilder : public IOpBuilder { // We still set the minimal supported opset to 1 as we couldn't // get the model opset version at this stage. virtual int GetMinSupportedOpSet(const Node& /* node */) const { return 1; } - virtual int GetMaxSupportedOpSet(const Node& /* node */) const { return 27; } + virtual int GetMaxSupportedOpSet(const Node& /* node */) const { return 28; } private: bool HasSupportedOpSet(const Node& node, const logging::Logger& logger) const; diff --git a/onnxruntime/core/providers/webnn/builders/impl/qdq_op_builder.cc b/onnxruntime/core/providers/webnn/builders/impl/qdq_op_builder.cc index e4e3a09391c3b..5b6ab6420ca45 100644 --- a/onnxruntime/core/providers/webnn/builders/impl/qdq_op_builder.cc +++ b/onnxruntime/core/providers/webnn/builders/impl/qdq_op_builder.cc @@ -25,6 +25,8 @@ class QDQOpBuilder : public BaseOpBuilder { const WebnnDeviceType /* device_type */, const logging::Logger& logger) const override; bool HasSupportedInputsImpl(const GraphViewer&, const Node& node, const emscripten::val& wnn_limits, const logging::Logger& logger) const override; + + int GetMaxSupportedOpSet(const Node& /* node */) const override { return 27; } }; Status QDQOpBuilder::AddToModelBuilderImpl(ModelBuilder& model_builder, diff --git a/onnxruntime/core/session/inference_session.cc b/onnxruntime/core/session/inference_session.cc index d126975211a7c..0065f3ed18bf7 100644 --- a/onnxruntime/core/session/inference_session.cc +++ b/onnxruntime/core/session/inference_session.cc @@ -1244,7 +1244,9 @@ common::Status InferenceSession::LoadOnnxModel(const PathString& model_uri) { const bool strict_shape_type_inference = session_options_.config_options.GetConfigOrDefault( kOrtSessionOptionsConfigStrictShapeTypeInference, "0") == "1"; - ModelOptions model_opts(true, strict_shape_type_inference, check_load_cancellation_fn_); + const bool allow_released_opsets_only = session_options_.config_options.GetConfigOrDefault( + kOrtSessionOptionsConfigStrictAllowReleasedOpsetsOnly, "1") == "1"; + ModelOptions model_opts(allow_released_opsets_only, strict_shape_type_inference, check_load_cancellation_fn_); // When set, the external initializers folder overrides the model's own directory as the // base for resolving external data. The model bytes are still read from model_uri. @@ -1340,6 +1342,8 @@ common::Status InferenceSession::Load(const void* model_data, int model_data_len const bool strict_shape_type_inference = session_options_.config_options.GetConfigOrDefault( kOrtSessionOptionsConfigStrictShapeTypeInference, "0") == "1"; + const bool allow_released_opsets_only = session_options_.config_options.GetConfigOrDefault( + kOrtSessionOptionsConfigStrictAllowReleasedOpsetsOnly, "1") == "1"; PathString external_data_model_path = GetExternalInitializersFolderModelPath(session_options_.config_options); if (!external_data_model_path.empty()) { @@ -1348,7 +1352,7 @@ common::Status InferenceSession::Load(const void* model_data, int model_data_len return onnxruntime::Model::Load(std::move(model_proto), model_location_, model, HasLocalSchema() ? &custom_schema_registries_ : nullptr, *session_logger_, - ModelOptions(true, strict_shape_type_inference, + ModelOptions(allow_released_opsets_only, strict_shape_type_inference, check_load_cancellation_fn_)); }; @@ -1405,6 +1409,8 @@ common::Status InferenceSession::LoadOnnxModel(ModelProto model_proto) { #endif const bool strict_shape_type_inference = session_options_.config_options.GetConfigOrDefault( kOrtSessionOptionsConfigStrictShapeTypeInference, "0") == "1"; + const bool allow_released_opsets_only = session_options_.config_options.GetConfigOrDefault( + kOrtSessionOptionsConfigStrictAllowReleasedOpsetsOnly, "1") == "1"; PathString external_data_model_path = GetExternalInitializersFolderModelPath(session_options_.config_options); if (!external_data_model_path.empty()) { @@ -1414,7 +1420,7 @@ common::Status InferenceSession::LoadOnnxModel(ModelProto model_proto) { // This call will move model_proto to the constructed model instance return onnxruntime::Model::Load(std::move(model_proto), model_location_, model, HasLocalSchema() ? &custom_schema_registries_ : nullptr, *session_logger_, - ModelOptions(true, strict_shape_type_inference, + ModelOptions(allow_released_opsets_only, strict_shape_type_inference, check_load_cancellation_fn_)); }; @@ -1516,12 +1522,15 @@ common::Status InferenceSession::Load(const OrtModel& model_editor_api_model) { const bool strict_shape_type_inference = session_options_.config_options.GetConfigOrDefault( kOrtSessionOptionsConfigStrictShapeTypeInference, "0") == "1"; + const bool allow_released_opsets_only = session_options_.config_options.GetConfigOrDefault( + kOrtSessionOptionsConfigStrictAllowReleasedOpsetsOnly, "1") == "1"; // need to go from unique_ptr to shared_ptr when moving into model_ std::unique_ptr tmp_model; ORT_RETURN_IF_ERROR(Model::LoadFromModelEditorApiModel(model_editor_api_model, HasLocalSchema() ? &custom_schema_registries_ : nullptr, - ModelOptions(true, strict_shape_type_inference, + ModelOptions(allow_released_opsets_only, + strict_shape_type_inference, check_load_cancellation_fn_), *session_logger_, tmp_model)); diff --git a/onnxruntime/python/backend/backend.py b/onnxruntime/python/backend/backend.py index 69be7a7657adf..5edac6df88fc0 100644 --- a/onnxruntime/python/backend/backend.py +++ b/onnxruntime/python/backend/backend.py @@ -135,6 +135,9 @@ def prepare(cls, model, device=None, **kwargs): return OnnxRuntimeBackendRep(model) elif isinstance(model, (str, bytes)): options = SessionOptions() + options.add_session_config_entry( + "session.allow_released_opsets_only", "1" if cls.allowReleasedOpsetsOnly else "0" + ) for k, v in kwargs.items(): if k in _ALLOWED_SESSION_OPTIONS: setattr(options, k, v) diff --git a/onnxruntime/python/backend/backend_rep.py b/onnxruntime/python/backend/backend_rep.py index 950ce417c6c2d..75b682fc9002e 100644 --- a/onnxruntime/python/backend/backend_rep.py +++ b/onnxruntime/python/backend/backend_rep.py @@ -6,6 +6,7 @@ Implements ONNX's backend API. """ +import numpy as np from onnx.backend.base import BackendRep from onnxruntime import RunOptions @@ -23,6 +24,10 @@ ) +def _normalize_numpy_scalar(value): + return np.asarray(value) if isinstance(value, np.generic) else value + + class OnnxRuntimeBackendRep(BackendRep): """ Wraps an :class:`onnxruntime.InferenceSession` to implement ONNX's @@ -61,7 +66,7 @@ def run(self, inputs, **kwargs): # type: (Any, **Any) -> Tuple[Any, ...] if isinstance(inputs, list): inps = {} for i, inp in enumerate(self._session.get_inputs()): - inps[inp.name] = inputs[i] + inps[inp.name] = _normalize_numpy_scalar(inputs[i]) outs = self._session.run(None, inps, options) if isinstance(outs, list): return outs @@ -72,5 +77,5 @@ def run(self, inputs, **kwargs): # type: (Any, **Any) -> Tuple[Any, ...] inp = self._session.get_inputs() if len(inp) != 1: raise RuntimeError(f"Model expect {len(inp)} inputs") - inps = {inp[0].name: inputs} + inps = {inp[0].name: _normalize_numpy_scalar(inputs)} return self._session.run(None, inps, options) diff --git a/onnxruntime/test/framework/tensorutils_test.cc b/onnxruntime/test/framework/tensorutils_test.cc index d238627f8e6df..5f5bcd5775a60 100644 --- a/onnxruntime/test/framework/tensorutils_test.cc +++ b/onnxruntime/test/framework/tensorutils_test.cc @@ -61,6 +61,8 @@ constexpr bool TensorProtoElementSizesAreConstexpr() { sizeof(uint8_t), // FLOAT8E8M0 sizeof(uint8_t), // UINT2 sizeof(uint8_t), // INT2 + sizeof(uint8_t), // FLOAT6E2M3 + sizeof(uint8_t), // FLOAT6E3M2 }; for (size_t index = 0; index < expected_sizes.size(); ++index) { diff --git a/onnxruntime/test/framework/type_info_test.cc b/onnxruntime/test/framework/type_info_test.cc index c6119778162e0..afc28d83d9c80 100644 --- a/onnxruntime/test/framework/type_info_test.cc +++ b/onnxruntime/test/framework/type_info_test.cc @@ -10,6 +10,7 @@ #include "core/framework/onnxruntime_map_type_info.h" #include "core/framework/onnxruntime_sequence_type_info.h" #include "core/framework/tensor_type_and_shape.h" +#include "core/framework/tensorprotoutils.h" #include "core/framework/onnxruntime_typeinfo.h" namespace onnxruntime { @@ -47,6 +48,8 @@ constexpr bool TensorElementTypeConversionIsConstexpr() { ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT8E8M0, ONNX_TENSOR_ELEMENT_DATA_TYPE_UINT2, ONNX_TENSOR_ELEMENT_DATA_TYPE_INT2, + ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E2M3, + ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E3M2, }; for (size_t index = 0; index < expected_types.size(); ++index) { @@ -65,6 +68,13 @@ constexpr bool TensorElementTypeConversionIsConstexpr() { static_assert(TensorElementTypeConversionIsConstexpr()); +TEST(TypeInfoTests, CApiElementTypeFromProtoTypeFloat6) { + EXPECT_EQ(utils::CApiElementTypeFromProtoType(ONNX_NAMESPACE::TensorProto_DataType_FLOAT6E2M3), + ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E2M3); + EXPECT_EQ(utils::CApiElementTypeFromProtoType(ONNX_NAMESPACE::TensorProto_DataType_FLOAT6E3M2), + ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT6E3M2); +} + TEST(TypeInfoTests, TensorProto) { mb::Type tensor_type = {1, 2, 3, 4}; diff --git a/onnxruntime/test/onnx/TestCase.cc b/onnxruntime/test/onnx/TestCase.cc index 86eb9623b4905..a2c54a6904ea8 100644 --- a/onnxruntime/test/onnx/TestCase.cc +++ b/onnxruntime/test/onnx/TestCase.cc @@ -958,6 +958,8 @@ std::unique_ptr> GetBrokenTests(const std::string& provider {"cast_UINT4_to_FLOAT", "Skipped until onnxruntime/cmake/external/onnx points to onnx 1.19 which should include @onnx/onnx/pull/7074"}, {"cast_UINT4_to_FLOAT16", "Skipped until onnxruntime/cmake/external/onnx points to onnx 1.19 which should include @onnx/onnx/pull/7074"}, {"cast_UINT4_to_UINT8", "Skipped until onnxruntime/cmake/external/onnx points to onnx 1.19 which should include @onnx/onnx/pull/7074"}, + {"reduce_max_empty_set_bool", + "ReduceMax does not support bool; mirrors onnx_backend_test_series_filters.jsonc global exclusion"}, // onnx#7959: the materialized node corpus now runs on the CPU EP via the // onnx_test_node_materialized ctest (previously the C++ node runner was QNN-only). ORT has // no per-channel-quantization ConvInteger support on any EP -- the CPU kernel throws and @@ -1091,8 +1093,6 @@ std::unique_ptr> GetBrokenTests(const std::string& provider {"resize_upsample_sizes_nearest_ceil_half_pixel", "Bad onnx test output. Needs test fix."}, {"resize_upsample_sizes_nearest_floor_align_corners", "Bad onnx test output. Needs test fix."}, {"resize_upsample_sizes_nearest_round_prefer_ceil_asymmetric", "Bad onnx test output. Needs test fix."}, - {"bitshift_right_uint16", "BitShift(11) uint16 support not enabled currently"}, - {"bitshift_left_uint16", "BitShift(11) uint16 support not enabled currently"}, {"maxunpool_export_with_output_shape", "Invalid output in ONNX test. See https://github.com/onnx/onnx/issues/2398"}, {"cntk_simple_seg", "Bad onnx test output caused by wrong SAME_UPPER/SAME_LOWER for ConvTranspose"}, diff --git a/onnxruntime/test/onnx/main.cc b/onnxruntime/test/onnx/main.cc index 5740a407c7ef6..1817a92c91e7f 100644 --- a/onnxruntime/test/onnx/main.cc +++ b/onnxruntime/test/onnx/main.cc @@ -908,6 +908,9 @@ select from 'TF8', 'TF16', 'UINT8', 'FLOAT', 'ITENSOR'. \n)"); ORT_TSTR("rotary_embedding_3d_input_expanded"), ORT_TSTR("rotary_embedding_expanded"), ORT_TSTR("rotary_embedding_interleaved_expanded"), + // QNN CPU returns uniform Softmax probabilities for the expanded function's all-negative-infinity rows, + // so its zero-row guard is not preserved. Keep the native Attention case enabled. + ORT_TSTR("attention_4d_causal_nonpad_negative_offset_structural_empty_expanded"), // QNN don't support fmod = 1 ORT_TSTR("mod_mixed_sign_float64"), ORT_TSTR("mod_mixed_sign_float32"), diff --git a/onnxruntime/test/onnx/tensorprotoutils.cc b/onnxruntime/test/onnx/tensorprotoutils.cc index 5c8d8f9fd8f23..c9cbc05b2f391 100644 --- a/onnxruntime/test/onnx/tensorprotoutils.cc +++ b/onnxruntime/test/onnx/tensorprotoutils.cc @@ -652,6 +652,8 @@ ONNXTensorElementDataType CApiElementTypeFromProtoType(int type) { CASE_TYPE(INT4) CASE_TYPE(UINT2) CASE_TYPE(INT2) + CASE_TYPE(FLOAT6E2M3) + CASE_TYPE(FLOAT6E3M2) default: return ONNX_TENSOR_ELEMENT_DATA_TYPE_UNDEFINED; } diff --git a/onnxruntime/test/optimizer/layout_transformation_potentially_added_ops_test.cc b/onnxruntime/test/optimizer/layout_transformation_potentially_added_ops_test.cc index c092c660633b1..c9fa32d569cf7 100644 --- a/onnxruntime/test/optimizer/layout_transformation_potentially_added_ops_test.cc +++ b/onnxruntime/test/optimizer/layout_transformation_potentially_added_ops_test.cc @@ -27,9 +27,15 @@ TEST(LayoutTransformationPotentiallyAddedOpsTests, OpsHaveLatestVersions) { if (prev_domain != it->domain || prev_op_type != it->op_type) { const auto* schema = schema_registry->GetSchema(std::string{it->op_type}, INT_MAX, std::string{it->domain}); ASSERT_NE(schema, nullptr); - EXPECT_EQ(schema->SinceVersion(), it->since_version) - << "A new version for op " << it->op_type << " (" << schema->SinceVersion() - << ") is available. Please update kLayoutTransformationPotentiallyAddedOps to include it."; + const bool is_staged_opset_28_qdq = + it->domain == kOnnxDomain && schema->SinceVersion() == 28 && + (it->op_type == "DequantizeLinear" || it->op_type == "QuantizeLinear"); + if (!is_staged_opset_28_qdq) { + EXPECT_EQ(schema->SinceVersion(), it->since_version) + << "A new version for op " << it->op_type << " (" << schema->SinceVersion() + << ") is available. Please update kLayoutTransformationPotentiallyAddedOps to include it."; + } + // TODO(#32372): Remove this exception when the opset-28 Q/DQ kernels land. prev_domain = it->domain; prev_op_type = it->op_type; } diff --git a/onnxruntime/test/optimizer/qdq_transformer_test.cc b/onnxruntime/test/optimizer/qdq_transformer_test.cc index 64f521d5c6825..aeb9c4eebd3bc 100644 --- a/onnxruntime/test/optimizer/qdq_transformer_test.cc +++ b/onnxruntime/test/optimizer/qdq_transformer_test.cc @@ -5636,7 +5636,10 @@ TEST(QDQTransformerTests, QDQPropagation_GH11605_Opset13) { // test removal of Q->DQ pairs by QDQFinalCleanupTransformer TEST(QDQTransformerTests, QDQFinalCleanupTransformerReportsIntentionalRemoval) { auto& logger = DefaultLoggingManager().DefaultLogger(); - Model model("QDQFinalCleanupRemovalTester", false, logger); + std::unordered_map domain_to_version; + domain_to_version[kOnnxDomain] = 25; + Model model("QDQFinalCleanupRemovalTester", false, ModelMetaData(), PathString(), + IOnnxRuntimeOpSchemaRegistryList(), domain_to_version, {}, logger); Graph& graph = model.MainGraph(); ModelTestBuilder builder(graph); diff --git a/onnxruntime/test/providers/cpu/controlflow/scan_test.cc b/onnxruntime/test/providers/cpu/controlflow/scan_test.cc index 4583b772bb30c..c35a45d590eda 100644 --- a/onnxruntime/test/providers/cpu/controlflow/scan_test.cc +++ b/onnxruntime/test/providers/cpu/controlflow/scan_test.cc @@ -1061,7 +1061,8 @@ TEST(Scan8, NumScanInputsExceedsVariadicInputs) { test.AddOutput("scan_output_2", {1, 2, 1}, {0.f, 0.f}); test.AddOutput("scan_output_3", {1, 2, 1}, {0.f, 0.f}); - test.Run(OpTester::ExpectResult::kExpectFailure, "Invalid 'num_scan_inputs' of 10. Value must be between 1 and 3", + test.Run(OpTester::ExpectResult::kExpectFailure, + "num_scan_inputs (10) plus the sequence_lens input cannot exceed the number of Scan inputs (4).", options.excluded_provider_types); } @@ -1165,7 +1166,7 @@ TEST(Scan9, NumScanInputsExceedsVariadicInputs) { // inferencing error wrapping a shape-inference error, rather than the kernel-level message produced // by the opset 8 test above. test.Run(OpTester::ExpectResult::kExpectFailure, - "[ShapeInferenceError] Mismatch between number of inferred and declared dimensions", + "num_scan_inputs (10) cannot exceed the number of Scan inputs (3).", options.excluded_provider_types); } #endif // !defined(ORT_NO_EXCEPTIONS) diff --git a/onnxruntime/test/providers/cpu/math/element_wise_ops_test.cc b/onnxruntime/test/providers/cpu/math/element_wise_ops_test.cc index f3cc72249b764..4348942186d17 100644 --- a/onnxruntime/test/providers/cpu/math/element_wise_ops_test.cc +++ b/onnxruntime/test/providers/cpu/math/element_wise_ops_test.cc @@ -15,6 +15,7 @@ #include #include #include +#include namespace onnxruntime { namespace test { @@ -4351,6 +4352,50 @@ TEST(ModOpTest, Fmod_float16_mixed_sign) { test.Run(OpTester::ExpectResult::kExpectSuccess, "", {kQnnExecutionProvider}); } +template +void TestFloorModFloatingPointOpset28(const std::vector& x, + const std::vector& y, + const std::vector& expected) { + OpTester test("Mod", 28); + test.AddAttribute("fmod", 0); + const std::vector shape{static_cast(x.size())}; + test.AddInput("X", shape, x); + test.AddInput("Y", shape, y); + test.AddOutput("Z", shape, expected); + test.Run(OpTester::ExpectResult::kExpectSuccess, "", + {kQnnExecutionProvider, kOpenVINOExecutionProvider}); +} + +TEST(ModOpTest, FloorMod_float_mixed_sign_opset28) { + TestFloorModFloatingPointOpset28( + {-4.3f, 7.2f, 5.0f, 4.3f, -7.2f, 8.0f}, + {2.1f, -3.4f, 8.0f, -2.1f, 3.4f, 5.0f}, + {2.0f, -3.0f, 5.0f, -2.0f, 3.0f, 3.0f}); +} + +TEST(ModOpTest, FloorMod_double_mixed_sign_opset28) { + TestFloorModFloatingPointOpset28( + {-4.3, 7.2, 5.0, 4.3, -7.2, 8.0}, + {2.1, -3.4, 8.0, -2.1, 3.4, 5.0}, + {2.0, -3.0, 5.0, -2.0, 3.0, 3.0}); +} + +TEST(ModOpTest, FloorMod_float16_mixed_sign_opset28) { + TestFloorModFloatingPointOpset28( + MakeMLFloat16({-4.3f, 7.2f, 5.0f, 4.3f, -7.2f, 8.0f}), + MakeMLFloat16({2.1f, -3.4f, 8.0f, -2.1f, 3.4f, 5.0f}), + MakeMLFloat16({2.0f, -3.0f, 5.0f, -2.0f, 3.0f, 3.0f})); +} + +TEST(ModOpTest, FloorMod_float_special_values_opset28) { + constexpr auto infinity = std::numeric_limits::infinity(); + constexpr auto nan = std::numeric_limits::quiet_NaN(); + TestFloorModFloatingPointOpset28( + {6.0f, -6.0f, 5.0f, -5.0f, infinity, 1.0f, nan, 1.0f}, + {-3.0f, 3.0f, infinity, -infinity, 2.0f, 0.0f, 2.0f, nan}, + {-0.0f, 0.0f, 5.0f, -5.0f, nan, nan, nan, nan}); +} + #if defined(USE_CUDA) TEST(ModOpTest, Fmod_bfloat16_mixed_sign) { OpTester test("Mod", 13); @@ -4652,6 +4697,93 @@ TEST(BitShiftOpTest, BroadcastXRight_Uint8) { test.Run(); } +template +void TestSignedBitShiftOpset28() { + constexpr T bit_width = static_cast(std::numeric_limits>::digits); + + { + OpTester test("BitShift", 28); + test.AddAttribute("direction", "LEFT"); + test.AddInput("X", {6}, {T{-8}, T{-3}, T{1}, std::numeric_limits::max(), T{-2}, T{2}}); + test.AddInput("Y", {6}, {T{1}, T{2}, T{3}, T{1}, T{-1}, bit_width}); + test.AddOutput("Z", {6}, {T{-16}, T{-12}, T{8}, T{-2}, T{0}, T{0}}); + test.Run(OpTester::ExpectResult::kExpectSuccess, "", {kOpenVINOExecutionProvider}); + } + + { + OpTester test("BitShift", 28); + test.AddAttribute("direction", "RIGHT"); + test.AddInput("X", {6}, {T{-8}, T{-3}, T{1}, T{-2}, T{-2}, T{2}}); + test.AddInput("Y", {6}, {T{1}, T{2}, T{3}, T{-1}, bit_width, bit_width}); + test.AddOutput("Z", {6}, {T{-4}, T{-1}, T{0}, T{-1}, T{-1}, T{0}}); + test.Run(OpTester::ExpectResult::kExpectSuccess, "", {kOpenVINOExecutionProvider}); + } + + { + OpTester test("BitShift", 28); + test.AddAttribute("direction", "RIGHT"); + test.AddInput("X", {1}, {T{-8}}); + test.AddInput("Y", {4}, {T{1}, T{2}, bit_width, T{-1}}); + test.AddOutput("Z", {4}, {T{-4}, T{-2}, T{-1}, T{-1}}); + test.Run(OpTester::ExpectResult::kExpectSuccess, "", {kOpenVINOExecutionProvider}); + } + + { + OpTester test("BitShift", 28); + test.AddAttribute("direction", "LEFT"); + test.AddInput("X", {2}, {std::numeric_limits::max(), T{-8}}); + test.AddInput("Y", {1}, {T{1}}); + test.AddOutput("Z", {2}, {T{-2}, T{-16}}); + test.Run(OpTester::ExpectResult::kExpectSuccess, "", {kOpenVINOExecutionProvider}); + } +} + +TEST(BitShiftOpTest, SignedInt8Opset28) { + TestSignedBitShiftOpset28(); +} + +TEST(BitShiftOpTest, SignedInt16Opset28) { + TestSignedBitShiftOpset28(); +} + +TEST(BitShiftOpTest, SignedInt32Opset28) { + TestSignedBitShiftOpset28(); +} + +TEST(BitShiftOpTest, SignedInt64Opset28) { + TestSignedBitShiftOpset28(); +} + +template +void TestUnsignedBitShiftRegistration(int opset) { + OpTester test("BitShift", opset); + test.AddAttribute("direction", "LEFT"); + test.AddInput("X", {2}, {T{1}, T{3}}); + test.AddInput("Y", {2}, {T{1}, T{2}}); + test.AddOutput("Z", {2}, {T{2}, T{12}}); + test.Run(); +} + +TEST(BitShiftOpTest, Uint16Opset11) { + TestUnsignedBitShiftRegistration(11); +} + +TEST(BitShiftOpTest, UnsignedTypesOpset28) { + TestUnsignedBitShiftRegistration(28); + TestUnsignedBitShiftRegistration(28); + TestUnsignedBitShiftRegistration(28); + TestUnsignedBitShiftRegistration(28); +} + +TEST(BitShiftOpTest, RightShiftByBitWidth_Uint64Opset28) { + OpTester test("BitShift", 28); + test.AddAttribute("direction", "RIGHT"); + test.AddInput("X", {4}, {1000, 255, 1, 42}); + test.AddInput("Y", {4}, {64, 64, 64, 64}); + test.AddOutput("Z", {4}, {0, 0, 0, 0}); + test.Run(); +} + // Test that shift amounts >= bit width produce 0 (not undefined behavior). // DirectML EP has the same hardware-level shift masking behavior, so skip these tests for DML. TEST(BitShiftOpTest, RightShiftByBitWidth_Uint64) { diff --git a/onnxruntime/test/providers/cpu/tensor/compress_op.test.cc b/onnxruntime/test/providers/cpu/tensor/compress_op.test.cc index 495fea5735b32..36c3f61a59630 100644 --- a/onnxruntime/test/providers/cpu/tensor/compress_op.test.cc +++ b/onnxruntime/test/providers/cpu/tensor/compress_op.test.cc @@ -1,6 +1,9 @@ // Copyright (c) Microsoft Corporation. All rights reserved. // Licensed under the MIT License. +#include +#include + #include "gtest/gtest.h" #include "test/providers/provider_test_utils.h" @@ -166,8 +169,13 @@ TEST(CompressTest, Compress_3dims_neg_axis) { // inputs to {0, 1}) and asserts the output is sized by truthiness. TEST(CompressTest, Compress_cuda_non_canonical_bool_condition) { // Build: output = Compress(input, condition, axis=0) - auto model = std::make_unique("compress_non_canonical_bool", false, - DefaultLoggingManager().DefaultLogger()); + const std::unordered_map domain_to_version{{kOnnxDomain, 27}}; + auto model = std::make_unique( + "compress_non_canonical_bool", false, ModelMetaData(), PathString(), IOnnxRuntimeOpSchemaRegistryList(), + domain_to_version, std::vector{}, DefaultLoggingManager().DefaultLogger()); + ASSERT_EQ(model->ToProto().opset_import_size(), 1); + EXPECT_EQ(model->ToProto().opset_import(0).domain(), kOnnxDomain); + EXPECT_EQ(model->ToProto().opset_import(0).version(), 27); onnxruntime::Graph& graph = model->MainGraph(); ONNX_NAMESPACE::TypeProto tensor_float; diff --git a/onnxruntime/test/providers/dml_onnx_tensor_wrapper_test.cc b/onnxruntime/test/providers/dml_onnx_tensor_wrapper_test.cc index c9d0d190e6ff9..6855b309f3e30 100644 --- a/onnxruntime/test/providers/dml_onnx_tensor_wrapper_test.cc +++ b/onnxruntime/test/providers/dml_onnx_tensor_wrapper_test.cc @@ -59,10 +59,9 @@ enum class ValueAttributePayload { // IMLOperatorAttributes1::GetTensorAttribute, which wraps the raw AttributeProto tensor in an // OnnxTensorWrapper. // -// Attribute tensors are not graph initializers, so none of the framework-level guards apply to them: -// Graph::ConvertInitializersIntoOrtValues never sees them, and ONNX shape inference for -// ConstantOfShape only reads the attribute's data_type (to type the output), never its data. That -// leaves OnnxTensorWrapper as the only place the payload can be validated. +// Attribute tensors are not graph initializers, so Graph::ConvertInitializersIntoOrtValues never +// sees them. Older ONNX versions also did not validate byte-aligned raw_data sizes in check_tensor, +// leaving OnnxTensorWrapper as the only place the payload could be validated. std::string BuildConstantOfShapeModel(ValueAttributePayload payload) { ONNX_NAMESPACE::ModelProto model; model.set_ir_version(ONNX_NAMESPACE::Version::IR_VERSION); @@ -113,8 +112,8 @@ std::string BuildConstantOfShapeModel(ValueAttributePayload payload) { return serialized; } -// Loads |serializedModel| into a session that has only the DML EP registered, and returns the status -// of Initialize(). |skipped| is set when no DirectML device is available. +// Loads |serializedModel| into a session that has only the DML EP registered, and returns the first +// failure from Load() or Initialize(). |skipped| is set when no DirectML device is available. // // Constant folding is disabled so the ConstantOfShape node survives graph optimization and is handed // to the DML EP. Without this, the CPU EP would evaluate the node during Level1 optimization and the @@ -166,15 +165,19 @@ TEST(DmlOnnxTensorWrapperTest, TruncatedRawDataAttributeTensorIsRejected) { GTEST_SKIP() << "DirectML execution provider is not available on this machine."; } - ASSERT_FALSE(status.IsOK()) << "Session initialization should reject a tensor whose declared shape " - "exceeds the data it carries."; - - // The detailed message raised inside the constructor does not survive: OnnxTensorWrapper is - // constructed behind the IMLOperatorAttributes1 COM boundary, whose ORT_CATCH_RETURN reduces the - // exception to a bare HRESULT before it is rethrown. 0x80070057 is E_INVALIDARG, so this asserts the - // model was rejected as malformed rather than failing for an unrelated reason. - EXPECT_NE(status.ErrorMessage().find("80070057"), std::string::npos) - << "Unexpected failure reason: " << status.ErrorMessage(); + ASSERT_FALSE(status.IsOK()) << "Model loading or session initialization should reject a tensor whose declared " + "shape exceeds the data it carries."; + + // ONNX 1.23 rejects the malformed attribute during Graph::Resolve. With older ONNX versions it + // reaches OnnxTensorWrapper, where the detailed constructor exception is reduced to E_INVALIDARG + // (0x80070057) at the IMLOperatorAttributes1 COM boundary. + const std::string& errorMessage = status.ErrorMessage(); + EXPECT_TRUE( + errorMessage.find("80070057") != std::string::npos || + errorMessage.find( + "raw_data size (1 bytes) is too small for the declared shape and type (8 bytes required)") != + std::string::npos) + << "Unexpected failure reason: " << errorMessage; } // The constructor's typed-field branch has the same defect in principle: UnpackTensor sizes the buffer diff --git a/onnxruntime/test/python/onnx_backend_test_series.py b/onnxruntime/test/python/onnx_backend_test_series.py index 62159721a2108..f12af65a6946f 100644 --- a/onnxruntime/test/python/onnx_backend_test_series.py +++ b/onnxruntime/test/python/onnx_backend_test_series.py @@ -17,6 +17,12 @@ import onnx.defs import onnxruntime.backend as backend # pylint: disable=consider-using-from-import +from onnxruntime.backend.backend import OnnxRuntimeBackend + +if os.getenv("ORT_BACKEND_TEST_ALLOW_UNRELEASED_OPSETS") == "1" and OnnxRuntimeBackend.allowReleasedOpsetsOnly: + raise RuntimeError( + "onnx_backend_test_series requires ALLOW_RELEASED_ONNX_OPSET_ONLY=0 before importing onnxruntime.backend." + ) pytest_plugins = ("onnx.backend.test.report",) diff --git a/onnxruntime/test/python/onnx_node_test_equivalence_test.py b/onnxruntime/test/python/onnx_node_test_equivalence_test.py index 1daa2f4d1eac7..32d4aace88251 100644 --- a/onnxruntime/test/python/onnx_node_test_equivalence_test.py +++ b/onnxruntime/test/python/onnx_node_test_equivalence_test.py @@ -30,9 +30,8 @@ backstop. Version parity: the materialized model.onnx opset/IR is stamped from the -compiled onnx schema registry, so this test asserts the installed onnx equals -the ``cmake/deps.txt`` pin (compared on the release base, so a pre-release rcN/.dev -wheel of the pinned tag is accepted). numpy parity is NOT asserted -- the core +compiled onnx schema registry, so this test asserts the installed onnx exactly +equals the ``cmake/deps.txt`` pin. numpy parity is NOT asserted -- the core treats numpy as an informational soft-warning, and recomputed float outputs that differ by within a small ULP band are accepted as Class A numpy-gen skew. """ @@ -62,14 +61,14 @@ def _repo_root() -> str: def _expected_onnx_version() -> str | None: - """Parse the onnx pin (archive tag) from cmake/deps.txt.""" + """Parse the ONNX_VERSION metadata paired with the archive pin in cmake/deps.txt.""" deps = os.path.join(_repo_root(), "cmake", "deps.txt") if not os.path.isfile(deps): return None with open(deps) as f: for line in f: - if line.startswith("onnx;"): - m = re.search(r"/tags/v([0-9]+\.[0-9]+\.[0-9]+)\.zip", line) + if line.startswith("# ONNX_VERSION="): + m = re.search(r"ONNX_VERSION=([0-9]+\.[0-9]+\.[0-9]+)", line) if m: return m.group(1) return None @@ -138,12 +137,11 @@ def test_materialized_equals_on_disk_oracle(self) -> None: except ImportError as exc: # pragma: no cover self.skipTest(f"onnx not importable: {exc}") - if expected_onnx and mat._release_base(onnx.__version__) != mat._release_base(expected_onnx): + if expected_onnx and onnx.__version__ != expected_onnx: self.skipTest( f"installed onnx=={onnx.__version__} != cmake/deps.txt pin " f"{expected_onnx}; cannot certify equivalence under version skew. " - f"Install the pinned onnx to run this check. (A pre-release rcN/.dev " - f"of the same release base is accepted.)" + "Install the pinned onnx to run this check." ) with tempfile.TemporaryDirectory(prefix="ort_node_mat_") as tmp: diff --git a/onnxruntime/test/python/onnxruntime_test_python_backend.py b/onnxruntime/test/python/onnxruntime_test_python_backend.py index bb83f6d36011f..24b61ec89203b 100644 --- a/onnxruntime/test/python/onnxruntime_test_python_backend.py +++ b/onnxruntime/test/python/onnxruntime_test_python_backend.py @@ -5,13 +5,17 @@ import os import tempfile import unittest +from types import SimpleNamespace import numpy as np from helper import get_name from numpy.testing import assert_allclose +from onnx import TensorProto, helper import onnxruntime as onnxrt import onnxruntime.backend as backend +from onnxruntime.backend.backend import OnnxRuntimeBackend +from onnxruntime.backend.backend_rep import OnnxRuntimeBackendRep class TestBackend(unittest.TestCase): @@ -23,6 +27,49 @@ def test_run_model(self): output_expected = np.array([[1.0, 4.0], [9.0, 16.0], [25.0, 36.0]], dtype=np.float32) np.testing.assert_allclose(res[0], output_expected, rtol=1e-05, atol=1e-08) + def test_prepare_model_bytes_allows_unreleased_opset_when_policy_disabled(self): + model = self._make_unreleased_opset_model() + + original_policy = OnnxRuntimeBackend.allowReleasedOpsetsOnly + try: + OnnxRuntimeBackend.allowReleasedOpsetsOnly = False + rep = backend.prepare(model.SerializeToString()) + finally: + OnnxRuntimeBackend.allowReleasedOpsetsOnly = original_policy + + result = rep.run(np.array([1.0], dtype=np.float32)) + np.testing.assert_array_equal(result[0], np.array([1.0], dtype=np.float32)) + + def test_prepare_model_path_allows_unreleased_opset_when_policy_disabled(self): + model = self._make_unreleased_opset_model() + with tempfile.TemporaryDirectory() as tmpdir: + model_path = os.path.join(tmpdir, "unreleased_opset.onnx") + with open(model_path, "wb") as model_file: + model_file.write(model.SerializeToString()) + + original_policy = OnnxRuntimeBackend.allowReleasedOpsetsOnly + try: + OnnxRuntimeBackend.allowReleasedOpsetsOnly = False + rep = backend.prepare(model_path) + finally: + OnnxRuntimeBackend.allowReleasedOpsetsOnly = original_policy + + result = rep.run(np.array([1.0], dtype=np.float32)) + np.testing.assert_array_equal(result[0], np.array([1.0], dtype=np.float32)) + + @staticmethod + def _make_unreleased_opset_model(): + model = helper.make_model( + helper.make_graph( + [helper.make_node("Identity", ["X"], ["Y"])], + "unreleased_opset", + [helper.make_tensor_value_info("X", TensorProto.FLOAT, [1])], + [helper.make_tensor_value_info("Y", TensorProto.FLOAT, [1])], + ), + opset_imports=[helper.make_opsetid("", 28)], + ) + return model + def test_allocation_plan_works_with_only_execute_path_to_fetches_option(self): """ (inp0) (inp1) @@ -66,6 +113,54 @@ def test_allocation_plan_works_with_only_execute_path_to_fetches_option(self): assert_allclose(session_run_results[0], -(inp0 - inp1)) +class TestBackendRepInputNormalization(unittest.TestCase): + class _Session: + def __init__(self, input_names): + self._inputs = [SimpleNamespace(name=name) for name in input_names] + self.received_inputs = None + + def get_inputs(self): + return self._inputs + + def run(self, _, inputs, __): + self.received_inputs = inputs + return [inputs] + + def test_run_normalizes_numpy_scalar_inputs_to_zero_dimensional_arrays(self): + session = self._Session(["first", "second"]) + rep = OnnxRuntimeBackendRep(session) + + rep.run([np.int32(3), np.float32(4.5)]) + + self.assertEqual(session.received_inputs["first"].shape, ()) + self.assertEqual(session.received_inputs["second"].shape, ()) + self.assertEqual(session.received_inputs["first"].dtype, np.dtype(np.int32)) + self.assertEqual(session.received_inputs["second"].dtype, np.dtype(np.float32)) + + def test_run_normalizes_single_numpy_scalar_to_zero_dimensional_array(self): + session = self._Session(["input"]) + rep = OnnxRuntimeBackendRep(session) + + rep.run(np.int64(7)) + + self.assertEqual(session.received_inputs["input"].shape, ()) + self.assertEqual(session.received_inputs["input"].dtype, np.dtype(np.int64)) + + def test_run_preserves_non_scalar_inputs(self): + values = ( + (np.array([1], dtype=np.int32),), + {"key": np.array([2], dtype=np.int32)}, + ) + for value in values: + with self.subTest(value=value): + session = self._Session(["input"]) + rep = OnnxRuntimeBackendRep(session) + + rep.run(value) + + self.assertIs(session.received_inputs["input"], value) + + class TestBackendKwargsAllowlist(unittest.TestCase): """Tests that the SessionOptions/RunOptions kwargs allowlist correctly blocks dangerous attributes and allows safe ones, preventing arbitrary file writes diff --git a/onnxruntime/test/python/requirements.txt b/onnxruntime/test/python/requirements.txt index 296ac83f3b6b5..48dff42186c8c 100644 --- a/onnxruntime/test/python/requirements.txt +++ b/onnxruntime/test/python/requirements.txt @@ -1,3 +1,3 @@ -onnx==1.22.0 +onnx==1.23.0 pytest onnx-ir diff --git a/onnxruntime/test/python/transformers/test_cuda_plugin_ep.py b/onnxruntime/test/python/transformers/test_cuda_plugin_ep.py index 95df52c3f6e61..670766f1c8fb0 100644 --- a/onnxruntime/test/python/transformers/test_cuda_plugin_ep.py +++ b/onnxruntime/test/python/transformers/test_cuda_plugin_ep.py @@ -30,8 +30,8 @@ EP_GRAPH_ASSIGNMENT_CONFIG_KEY = "session.record_ep_graph_assignment_info" NO_CUDNN_PLUGIN_TEST = os.getenv("ORT_TEST_CUDA_PLUGIN_NO_CUDNN", "").upper() in {"1", "ON", "TRUE", "YES"} requires_cudnn = unittest.skipIf(NO_CUDNN_PLUGIN_TEST, "test requires cuDNN-backed CUDA plugin kernels") -# Use the latest released ai.onnx opset so the model builders stay current as ONNX releases new opsets. -DEFAULT_ONNX_OPSET = max(v for (d, v) in helper.OP_SET_ID_VERSION_MAP if d == "ai.onnx") +# CUDA plugin runtime currently supports released ai.onnx opsets through 27. +DEFAULT_ONNX_OPSET = 27 def _make_released_opset_model(graph, producer_name="onnx-example"): diff --git a/onnxruntime/test/python/transformers/test_moe_cuda.py b/onnxruntime/test/python/transformers/test_moe_cuda.py index 5c92b5e197e4f..11731f295fad6 100644 --- a/onnxruntime/test/python/transformers/test_moe_cuda.py +++ b/onnxruntime/test/python/transformers/test_moe_cuda.py @@ -1734,7 +1734,11 @@ def create_sparse_mixer_onnx_graph( ], ) - return helper.make_model(graph, producer_name="MoE_Model") + return helper.make_model( + graph, + producer_name="MoE_Model", + opset_imports=[helper.make_opsetid("", 27), helper.make_opsetid("com.microsoft", 1)], + ) class TestSparseMixer(unittest.TestCase): @@ -1777,6 +1781,8 @@ def test_sparse_mixer_functional(self, onnx_dtype): fc2_bias, onnx_dtype, ) + opset_imports = {opset_import.domain: opset_import.version for opset_import in onnx_model.opset_import} + self.assertEqual(opset_imports, {"": 27, "com.microsoft": 1}) sess_options = onnxruntime.SessionOptions() sess = onnxruntime.InferenceSession(onnx_model.SerializeToString(), sess_options, providers=get_ort_provider()) diff --git a/onnxruntime/test/testdata/onnx_backend_test_series_filters.jsonc b/onnxruntime/test/testdata/onnx_backend_test_series_filters.jsonc index a10eb97496186..cf40b161dd008 100644 --- a/onnxruntime/test/testdata/onnx_backend_test_series_filters.jsonc +++ b/onnxruntime/test/testdata/onnx_backend_test_series_filters.jsonc @@ -55,6 +55,18 @@ "^test_attention_(23|24)_.*qk_matmul_output_mode3", // #28994 CUDA NOT_IMPLEMENTED mode>kQK (attention.cc:1477); #27712 to support // TODO: support qk_matmul_output modes beyond kQK in Attention-cuda (see issue #27712) // Tests combining qk_matmul with softcap need unfused-path qk_matmul support (deferred). + "^test_attention_3d_local_window_cuda$", // exits with 0xDEAD in the CUDA backend test + "^test_attention_bidirectional_window_cuda$", // exits with 0xDEAD in the CUDA backend test + "^test_attention_local_window_cuda$", // exits with 0xDEAD in the CUDA backend test + "^test_attention_local_window_ext_cache_float16_mask_cuda$", // exits with 0xDEAD in the CUDA backend test + "^test_attention_local_window_ext_cache_rank2_mask_cuda$", // exits with 0xDEAD in the CUDA backend test + "^test_attention_local_window_ext_cache_rank3_head_mask_cuda$", // exits with 0xDEAD in the CUDA backend test + "^test_attention_local_window_ext_cache_rank4_batch_mask_cuda$", // exits with 0xDEAD in the CUDA backend test + "^test_attention_local_window_gqa_rank4_mask_cuda$", // exits with an access violation in the CUDA backend test + "^test_attention_local_window_gqa_rank4_mask_expanded_cuda$", // expanded Attention output mismatch + "^test_attention_local_window_rank1_boolean_mask_cuda$", // exits with an access violation in the CUDA backend test + "^test_attention_local_window_with_past_cuda$", // exits with an access violation in the CUDA backend test + "^test_attention_4d_causal_nonpad_negative_offset_structural_empty_expanded_cuda$", // expanded Attention output mismatch "^test_attention_3d_with_past_and_present_qk_matmul_softcap_cuda", // qk_matmul modes beyond kQK not supported "^test_attention_4d_with_qk_matmul_softcap_cuda", // qk_matmul modes beyond kQK not supported "^test_attention_3d_with_past_and_present_qk_matmul_bias_cuda", // QK matmul + bias not supported in Attention-cuda @@ -84,6 +96,209 @@ "^test_castlike_UINT2_to*", // ORT does not support uint2 "^test_cast_INT2_to*", // ORT does not support int2 "^test_cast_UINT2_to*", // ORT does not support uint2 + // TODO(#32378): Remove the exact BF16 Attention failures when the function dependencies support BF16. + "^test_attention_3d_causal_bf16_cpu$", + "^test_attention_3d_causal_bf16_expanded_cpu$", + "^test_attention_4d_attn_mask_causal_bf16_cpu$", + "^test_attention_4d_attn_mask_causal_bf16_expanded_cpu$", + "^test_attention_4d_causal_bf16_cpu$", + "^test_attention_4d_causal_bf16_expanded_cpu$", + "^test_attention_4d_causal_padded_kv_bf16_cpu$", + "^test_attention_4d_causal_padded_kv_bf16_expanded_cpu$", + "^test_attention_4d_padded_kv_bf16_cpu$", + "^test_attention_4d_padded_kv_bf16_expanded_cpu$", + "^test_attention_3d_causal_bf16_cuda$", + "^test_attention_3d_causal_bf16_expanded_cuda$", + "^test_attention_4d_attn_mask_causal_bf16_cuda$", + "^test_attention_4d_attn_mask_causal_bf16_expanded_cuda$", + "^test_attention_4d_causal_bf16_cuda$", + "^test_attention_4d_causal_bf16_expanded_cuda$", + "^test_attention_4d_causal_padded_kv_bf16_cuda$", + "^test_attention_4d_causal_padded_kv_bf16_expanded_cuda$", + "^test_attention_4d_padded_kv_bf16_cuda$", + "^test_attention_4d_padded_kv_bf16_expanded_cuda$", + "^test_attention_local_window_ext_cache_float16_mask_cpu$", + // TODO: Remove the exact Cast(28) failures when CPU Cast(28) is available in all builds. + "^test_cast_DOUBLE_to_FLOAT_cpu$", + "^test_celu_bfloat16_cpu$", + "^test_celu_bfloat16_expanded_cpu$", + "^test_celu_cpu$", + "^test_celu_expanded_cpu$", + "^test_celu_float16_cpu$", + "^test_celu_float16_expanded_cpu$", + // TODO(#32381): Remove the exact log-reduction failures when their Cast(28) dependency is available. + "^test_reduce_log_sum_exp_default_axes_keepdims_example_cpu$", + "^test_reduce_log_sum_exp_default_axes_keepdims_example_expanded_cpu$", + "^test_reduce_log_sum_exp_default_axes_keepdims_random_cpu$", + "^test_reduce_log_sum_exp_default_axes_keepdims_random_expanded_cpu$", + "^test_reduce_log_sum_exp_do_not_keepdims_example_cpu$", + "^test_reduce_log_sum_exp_do_not_keepdims_example_expanded_cpu$", + "^test_reduce_log_sum_exp_do_not_keepdims_random_cpu$", + "^test_reduce_log_sum_exp_do_not_keepdims_random_expanded_cpu$", + "^test_reduce_log_sum_exp_empty_set_cpu$", + "^test_reduce_log_sum_exp_empty_set_expanded_cpu$", + "^test_reduce_log_sum_exp_keepdims_example_cpu$", + "^test_reduce_log_sum_exp_keepdims_example_expanded_cpu$", + "^test_reduce_log_sum_exp_keepdims_random_cpu$", + "^test_reduce_log_sum_exp_keepdims_random_expanded_cpu$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_example_cpu$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_example_expanded_cpu$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_random_cpu$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_random_expanded_cpu$", + // TODO: Remove the exact Compress(28) failures when the existing kernel supports opset 28. + "^test_compress_0_cpu$", + "^test_compress_1_cpu$", + "^test_compress_bfloat16_cpu$", + "^test_compress_default_axis_cpu$", + "^test_compress_negative_axis_cpu$", + // TODO: Remove the exact Einsum(28) failures when the existing kernel supports opset 28. + "^test_einsum_batch_diagonal_cpu$", + "^test_einsum_batch_matmul_bfloat16_cpu$", + "^test_einsum_batch_matmul_cpu$", + "^test_einsum_inner_prod_cpu$", + "^test_einsum_scalar_cpu$", + "^test_einsum_sum_bfloat16_cpu$", + "^test_einsum_sum_cpu$", + "^test_einsum_transpose_bfloat16_cpu$", + "^test_einsum_transpose_cpu$", + // TODO: Remove the exact OneHot(28) failures when the existing kernel supports opset 28. + "^test_onehot_negative_indices_cpu$", + "^test_onehot_out_of_range_indices_cpu$", + "^test_onehot_with_axis_cpu$", + "^test_onehot_with_bfloat16_values_cpu$", + "^test_onehot_with_negative_axis_cpu$", + "^test_onehot_without_axis_cpu$", + // TODO: Remove this test when ReduceMax(28) supports an empty bool input. + "^test_reduce_max_empty_set_bool_cpu$", + "^test_reduce_max_empty_set_bool_cuda$", + // TODO(#32381): Remove the exact Swish(24) function-dependency failures when SwiGLU is supported. + "^test_swiglu_alpha_cpu$", + "^test_swiglu_alpha_expanded_cpu$", + "^test_swiglu_cpu$", + "^test_swiglu_expanded_cpu$", + "^test_swiglu_float16_cpu$", + "^test_swiglu_float16_expanded_cpu$", + // TODO: Remove the exact Unique(28) failures when the existing kernel supports opset 28. + "^test_unique_bfloat16_sorted_without_axis_cpu$", + "^test_unique_length_1_cpu$", + "^test_unique_not_sorted_without_axis_cpu$", + "^test_unique_sorted_with_axis_3d_cpu$", + "^test_unique_sorted_with_axis_cpu$", + "^test_unique_sorted_with_negative_axis_cpu$", + "^test_unique_sorted_without_axis_cpu$", + // TODO: Remove these exact CUDA skips when fallback nodes have providers and Swish(24) has a CPU kernel. + "^test_cast_DOUBLE_to_FLOAT_cuda$", + "^test_celu_bfloat16_cuda$", + "^test_celu_bfloat16_expanded_cuda$", + "^test_celu_cuda$", + "^test_celu_expanded_cuda$", + "^test_celu_float16_cuda$", + "^test_celu_float16_expanded_cuda$", + "^test_compress_0_cuda$", + "^test_compress_1_cuda$", + "^test_compress_bfloat16_cuda$", + "^test_compress_default_axis_cuda$", + "^test_compress_negative_axis_cuda$", + "^test_einsum_batch_diagonal_cuda$", + "^test_einsum_batch_matmul_bfloat16_cuda$", + "^test_einsum_batch_matmul_cuda$", + "^test_einsum_inner_prod_cuda$", + "^test_einsum_scalar_cuda$", + "^test_einsum_sum_bfloat16_cuda$", + "^test_einsum_sum_cuda$", + "^test_einsum_transpose_bfloat16_cuda$", + "^test_einsum_transpose_cuda$", + "^test_onehot_negative_indices_cuda$", + "^test_onehot_out_of_range_indices_cuda$", + "^test_onehot_with_axis_cuda$", + "^test_onehot_with_bfloat16_values_cuda$", + "^test_onehot_with_negative_axis_cuda$", + "^test_onehot_without_axis_cuda$", + "^test_reduce_log_sum_exp_default_axes_keepdims_example_cuda$", + "^test_reduce_log_sum_exp_default_axes_keepdims_example_expanded_cuda$", + "^test_reduce_log_sum_exp_default_axes_keepdims_random_cuda$", + "^test_reduce_log_sum_exp_default_axes_keepdims_random_expanded_cuda$", + "^test_reduce_log_sum_exp_do_not_keepdims_example_cuda$", + "^test_reduce_log_sum_exp_do_not_keepdims_example_expanded_cuda$", + "^test_reduce_log_sum_exp_do_not_keepdims_random_cuda$", + "^test_reduce_log_sum_exp_do_not_keepdims_random_expanded_cuda$", + "^test_reduce_log_sum_exp_keepdims_example_cuda$", + "^test_reduce_log_sum_exp_keepdims_example_expanded_cuda$", + "^test_reduce_log_sum_exp_keepdims_random_cuda$", + "^test_reduce_log_sum_exp_keepdims_random_expanded_cuda$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_example_cuda$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_example_expanded_cuda$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_random_cuda$", + "^test_reduce_log_sum_exp_negative_axes_keepdims_random_expanded_cuda$", + "^test_swiglu_alpha_cuda$", + "^test_swiglu_alpha_expanded_cuda$", + "^test_swiglu_cuda$", + "^test_swiglu_expanded_cuda$", + "^test_swiglu_float16_cuda$", + "^test_swiglu_float16_expanded_cuda$", + "^test_unique_bfloat16_sorted_without_axis_cuda$", + "^test_unique_length_1_cuda$", + "^test_unique_not_sorted_without_axis_cuda$", + "^test_unique_sorted_with_axis_3d_cuda$", + "^test_unique_sorted_with_axis_cuda$", + "^test_unique_sorted_with_negative_axis_cuda$", + "^test_unique_sorted_without_axis_cuda$", + // TODO(#32361): Remove these exact skips when the opset-28 BitShift kernels land. + ["opset28", "^test_bitshift_left_int16(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int32(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int32_negative_shift(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int32_overflow(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int32_shift_ge_width(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int64(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int8(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int8_negative_shift(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int8_overflow(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_int8_shift_ge_width(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_uint32(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_uint64(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_left_uint8(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int16(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int32(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int32_negative_input(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int32_negative_shift(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int32_shift_ge_width(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int64(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int8(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int8_negative_input(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int8_negative_shift(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_int8_shift_ge_width(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_uint32(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_uint64(_(cpu|cuda))?$"], + ["opset28", "^test_bitshift_right_uint8(_(cpu|cuda))?$"], + "^test_mod_mixed_sign_int16_cuda$", + "^test_mod_mixed_sign_int8_cuda$", + "^test_mod_uint16_cuda$", + "^test_mod_uint8_cuda$", + // TODO(#32372): Remove these exact skips when the opset-28 Q/DQ kernels land. + "^test_dequantizelinear(_(cpu|cuda))?$", + "^test_dequantizelinear_axis(_(cpu|cuda))?$", + "^test_dequantizelinear_e4m3fn(_(cpu|cuda))?$", + "^test_dequantizelinear_e4m3fn_float16(_(cpu|cuda))?$", + "^test_dequantizelinear_e4m3fn_zero_point(_(cpu|cuda))?$", + "^test_dequantizelinear_e5m2(_(cpu|cuda))?$", + "^test_dequantizelinear_float4e2m1(_(cpu|cuda))?$", + "^test_dequantizelinear_int16(_(cpu|cuda))?$", + "^test_dequantizelinear_int2(_(cpu|cuda))?$", + "^test_dequantizelinear_uint16(_(cpu|cuda))?$", + "^test_dequantizelinear_uint2(_(cpu|cuda))?$", + "^test_quantizelinear(_(cpu|cuda))?$", + "^test_quantizelinear_axis(_(cpu|cuda))?$", + "^test_quantizelinear_e4m3fn(_(cpu|cuda))?$", + "^test_quantizelinear_e5m2(_(cpu|cuda))?$", + "^test_quantizelinear_float4e2m1(_(cpu|cuda))?$", + "^test_quantizelinear_int16(_(cpu|cuda))?$", + "^test_quantizelinear_int2(_(cpu|cuda))?$", + "^test_quantizelinear_uint16(_(cpu|cuda))?$", + "^test_quantizelinear_uint2(_(cpu|cuda))?$", + // TODO(#32378): Remove these exact skips when opset-28 ReverseSequence support lands. + ["opset28", "^test_reversesequence_batch(_(cpu|cuda))?$"], + ["opset28", "^test_reversesequence_bfloat16(_(cpu|cuda))?$"], + ["opset28", "^test_reversesequence_time(_(cpu|cuda))?$"], "^test_convinteger_with_padding*", // ORT does not support per-channel quantization for ConvInteger "^test_cast_e8m0_*", // ORT does not support float8e8m0 "^test_batchnorm_epsilon_training_mode", diff --git a/onnxruntime/test/testdata/onnx_backend_test_series_overrides.jsonc b/onnxruntime/test/testdata/onnx_backend_test_series_overrides.jsonc index f4da6b0d37a47..b56b0e5a97720 100644 --- a/onnxruntime/test/testdata/onnx_backend_test_series_overrides.jsonc +++ b/onnxruntime/test/testdata/onnx_backend_test_series_overrides.jsonc @@ -21,5 +21,8 @@ }, // Key: str, the name of the test as defined by ONNX without any device suffix. // Val: float, max relative difference between expected and actual. - "rtol_overrides": {} + "rtol_overrides": { + "test_attention_4d_causal_fp16": 0.002, + "test_attention_4d_causal_fp16_expanded": 0.002 + } } diff --git a/onnxruntime/test/unittest_util/graph_transform_test_builder.cc b/onnxruntime/test/unittest_util/graph_transform_test_builder.cc index 325f820343091..4e67113a4d76a 100644 --- a/onnxruntime/test/unittest_util/graph_transform_test_builder.cc +++ b/onnxruntime/test/unittest_util/graph_transform_test_builder.cc @@ -169,9 +169,7 @@ void TransformerTester(const std::function& buil SessionOptions session_options; session_options.graph_optimization_level = transformer ? baseline_level : level; // Mirror the model build's shape/type-inference strictness onto the session so that - // loading the serialized model applies the same options. (allow_released_opsets_only is - // passed explicitly to Load below, since the byte/proto Load overloads don't read it - // from the config options.) + // loading the serialized model applies the same options. ASSERT_STATUS_OK(session_options.config_options.AddConfigEntry( kOrtSessionOptionsConfigStrictShapeTypeInference, model_options.strict_shape_type_inference ? "1" : "0")); @@ -187,8 +185,7 @@ void TransformerTester(const std::function& buil ASSERT_STATUS_OK(session.RegisterExecutionProvider(ep_shared)); } - // Load via an istream so the under-development opset flag can be honored: the istream Load - // overload takes allow_released_opsets_only explicitly (the byte/proto overloads hardcode it). + // Load via an istream so the under-development opset flag can be passed explicitly. // This lets models built at an unreleased ONNX opset (e.g. opset 27 in ONNX 1.22) load on // strict (ALLOW_RELEASED_ONNX_OPSET_ONLY default) CI legs. std::istringstream model_istream(model_data); diff --git a/onnxruntime/test/unittest_util/op_tester.cc b/onnxruntime/test/unittest_util/op_tester.cc index b633989866277..2b1f4d87df2e1 100644 --- a/onnxruntime/test/unittest_util/op_tester.cc +++ b/onnxruntime/test/unittest_util/op_tester.cc @@ -40,6 +40,12 @@ onnxruntime::Model& OpTester::BuildModel(const std::unordered_map domain_to_version(extra_domain_to_version); const auto& domain = Domain(); + if (domain != kOnnxDomain && domain_to_version.count(kOnnxDomain) == 0) { + const auto& last_released_versions = + ONNX_NAMESPACE::OpSchemaRegistry::DomainToVersionRange::Instance().LastReleaseVersionMap(); + domain_to_version.insert({kOnnxDomain, last_released_versions.at(kOnnxDomain)}); + } + if (domain_to_version.count(domain) == 0) { domain_to_version.insert({domain, Opset()}); } else { diff --git a/requirements-materialize-onnx-node-tests.txt b/requirements-materialize-onnx-node-tests.txt index ee46a7f7ea7ed..35abc6a0983c6 100644 --- a/requirements-materialize-onnx-node-tests.txt +++ b/requirements-materialize-onnx-node-tests.txt @@ -10,6 +10,6 @@ # PIN-LOCK: these values MUST stay in lockstep with the CMake gate (cmake/onnxruntime_unittests.cmake, # which HARD-FATALs when the installed versions differ from its expected onnx/numpy pins) and with # the Python MIN pins. Bump here and in cmake/deps.txt + the numpy MIN pins together. -onnx==1.22.0 +onnx==1.23.0 numpy==2.2.6; python_version < "3.11" numpy==2.4.2; python_version >= "3.11" diff --git a/tools/ci_build/build.py b/tools/ci_build/build.py index fe1bfa3f637d8..8a201cb10356d 100644 --- a/tools/ci_build/build.py +++ b/tools/ci_build/build.py @@ -149,6 +149,13 @@ def run_subprocess( return run(*args, cwd=cwd, capture_stdout=capture_stdout, shell=shell, env=my_env) +def get_onnx_backend_test_environment(_use_cuda): + return { + "ALLOW_RELEASED_ONNX_OPSET_ONLY": "0", + "ORT_BACKEND_TEST_ALLOW_UNRELEASED_OPSETS": "1", + } + + def update_submodules(source_dir): run_subprocess(["git", "submodule", "sync", "--recursive"], cwd=source_dir) run_subprocess(["git", "submodule", "update", "--init", "--recursive"], cwd=source_dir) @@ -2066,7 +2073,13 @@ def run_onnxruntime_tests(args, source_dir, ctest_path, build_dir, configs): if not args.skip_onnx_tests: run_subprocess([os.path.join(cwd, "onnx_test_runner"), "test_models"], cwd=cwd) if config != "Debug": - run_subprocess([sys.executable, "onnx_backend_test_series.py"], cwd=cwd, dll_path=dll_path) + # Set the opset policy explicitly so the child process does not inherit the CI default. + run_subprocess( + [sys.executable, "onnx_backend_test_series.py"], + cwd=cwd, + dll_path=dll_path, + env=get_onnx_backend_test_environment(args.use_cuda), + ) if not args.skip_keras_test: try: diff --git a/tools/ci_build/github/linux/build_linux_python_package.sh b/tools/ci_build/github/linux/build_linux_python_package.sh index abd3926182da7..a5c4aa3c45fff 100755 --- a/tools/ci_build/github/linux/build_linux_python_package.sh +++ b/tools/ci_build/github/linux/build_linux_python_package.sh @@ -118,6 +118,10 @@ fi export ONNX_ML=1 export CMAKE_ARGS="-DONNX_GEN_PB_TYPE_STUBS=ON -DONNX_WERROR=OFF" +FREE_THREADED_REQUIREMENTS=$(mktemp) +grep -Ev '^(onnx|onnx-ir|onnxscript)==' \ + /onnxruntime_src/tools/ci_build/github/linux/python/requirements.txt > "${FREE_THREADED_REQUIREMENTS}" +trap 'rm -f "${FREE_THREADED_REQUIREMENTS}"' EXIT for PYTHON_EXE in "${PYTHON_EXES[@]}" do @@ -136,6 +140,10 @@ do python3_dir=$(dirname "$PYTHON_EXE") ls "$python3_dir" PIP_REQUIREMENTS=(-r /onnxruntime_src/tools/ci_build/github/linux/python/requirements.txt) + if [[ "${PYTHON_EXE}" == */cp3??-cp3??t/* ]]; then + # ONNX does not publish free-threaded wheels, and ONNX-dependent tooling would resolve its unsupported source build. + PIP_REQUIREMENTS=(-r "${FREE_THREADED_REQUIREMENTS}") + fi if [[ "${PYTHON_EXE}" == */cp313-cp313t/* ]]; then # mypy 1.19+ dependencies do not support free-threaded CPython 3.13. PIP_REQUIREMENTS+=("mypy<1.19") diff --git a/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/install_deps.sh b/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/install_deps.sh index 309f0ad8d8b51..7458fc7b8d069 100755 --- a/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/install_deps.sh +++ b/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/install_deps.sh @@ -9,10 +9,17 @@ fi popd export ONNX_ML=1 export CMAKE_ARGS="-DONNX_GEN_PB_TYPE_STUBS=ON -DONNX_WERROR=OFF" +FREE_THREADED_REQUIREMENTS=$(mktemp) +grep -Ev '^(onnx|onnx-ir|onnxscript)==' requirements.txt > "${FREE_THREADED_REQUIREMENTS}" +trap 'rm -f "${FREE_THREADED_REQUIREMENTS}"' EXIT for PYTHON_EXE in "${PYTHON_EXES[@]}" do PIP_REQUIREMENTS=(-r requirements.txt) + if [[ "${PYTHON_EXE}" == */cp3??-cp3??t/* ]]; then + # ONNX does not publish free-threaded wheels, and ONNX-dependent tooling would resolve its unsupported source build. + PIP_REQUIREMENTS=(-r "${FREE_THREADED_REQUIREMENTS}") + fi if [[ "${PYTHON_EXE}" == */cp313-cp313t/* ]]; then # mypy 1.19+ dependencies do not support free-threaded CPython 3.13. PIP_REQUIREMENTS+=("mypy<1.19") diff --git a/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/requirements.txt b/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/requirements.txt index e5b4b9b2dbd94..695ecb7cffdf0 100644 --- a/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/requirements.txt +++ b/tools/ci_build/github/linux/docker/inference/aarch64/python/cpu/scripts/requirements.txt @@ -7,4 +7,4 @@ wheel protobuf==6.33.6 sympy==1.14 flatbuffers -onnx==1.22.0 +onnx==1.23.0 diff --git a/tools/ci_build/github/linux/docker/scripts/lort/requirements.txt b/tools/ci_build/github/linux/docker/scripts/lort/requirements.txt index 48a4c95a3e7da..fc8e614a1dd05 100644 --- a/tools/ci_build/github/linux/docker/scripts/lort/requirements.txt +++ b/tools/ci_build/github/linux/docker/scripts/lort/requirements.txt @@ -3,7 +3,7 @@ beartype==0.15.0 flatbuffers cerberus h5py -onnx==1.22.0 +onnx==1.23.0 # Python dependencies required for pytorch development astunparse expecttest!=0.2.0 diff --git a/tools/ci_build/github/linux/docker/scripts/manylinux/requirements.txt b/tools/ci_build/github/linux/docker/scripts/manylinux/requirements.txt index 5b0a682d9d419..f7f377dee752b 100644 --- a/tools/ci_build/github/linux/docker/scripts/manylinux/requirements.txt +++ b/tools/ci_build/github/linux/docker/scripts/manylinux/requirements.txt @@ -9,4 +9,4 @@ sympy==1.14 flatbuffers neural-compressor>=2.2.1 triton==3.5.0 -onnx==1.22.0 +onnx==1.23.0 diff --git a/tools/ci_build/github/linux/docker/scripts/requirements.txt b/tools/ci_build/github/linux/docker/scripts/requirements.txt index 462de90e67c3f..42e559c7a8a4e 100644 --- a/tools/ci_build/github/linux/docker/scripts/requirements.txt +++ b/tools/ci_build/github/linux/docker/scripts/requirements.txt @@ -12,4 +12,4 @@ protobuf==6.33.6 packaging onnxscript==0.6.2 onnx-ir==0.1.16 -onnx==1.22.0 +onnx==1.23.0 diff --git a/tools/ci_build/github/linux/python/requirements.txt b/tools/ci_build/github/linux/python/requirements.txt index 4730f767c9b50..3206255095611 100644 --- a/tools/ci_build/github/linux/python/requirements.txt +++ b/tools/ci_build/github/linux/python/requirements.txt @@ -12,4 +12,4 @@ onnxscript==0.6.2 onnx-ir==0.1.16 jinja2 markupsafe -onnx==1.22.0 +onnx==1.23.0 diff --git a/tools/ci_build/github/windows/python/requirements.txt b/tools/ci_build/github/windows/python/requirements.txt index ef80b8af11b05..7a001217c950a 100644 --- a/tools/ci_build/github/windows/python/requirements.txt +++ b/tools/ci_build/github/windows/python/requirements.txt @@ -14,4 +14,4 @@ jinja2 markupsafe semver packaging -onnx==1.22.0 +onnx==1.23.0 diff --git a/tools/ci_build/test_build_args.py b/tools/ci_build/test_build_args.py index a98eaaaed1974..b2c93ed54248f 100644 --- a/tools/ci_build/test_build_args.py +++ b/tools/ci_build/test_build_args.py @@ -2,6 +2,7 @@ # Copyright (c) Microsoft Corporation. All rights reserved. # Licensed under the MIT License. +import importlib import sys import unittest from pathlib import Path @@ -10,7 +11,13 @@ sys.path.insert(0, str(Path(__file__).resolve().parents[1] / "python")) sys.path.insert(0, str(Path(__file__).resolve().parent)) -import build_args +build = importlib.import_module("build") +build_args = importlib.import_module("build_args") + +_UNRELEASED_OPSET_ENVIRONMENT = { + "ALLOW_RELEASED_ONNX_OPSET_ONLY": "0", + "ORT_BACKEND_TEST_ALLOW_UNRELEASED_OPSETS": "1", +} class BuildArgsTest(unittest.TestCase): @@ -65,5 +72,25 @@ def test_acl_deprecation_warning_not_emitted_without_use_acl(self): warn.assert_not_called() +class OnnxBackendTestEnvironmentTest(unittest.TestCase): + def test_cpu_and_cuda_enable_unreleased_opsets_by_default(self): + with mock.patch.dict(build.os.environ, {}, clear=True): + for use_cuda in (False, True): + with self.subTest(use_cuda=use_cuda): + self.assertEqual( + build.get_onnx_backend_test_environment(use_cuda), + _UNRELEASED_OPSET_ENVIRONMENT, + ) + + def test_cpu_and_cuda_override_explicit_parent_strict_opset_mode(self): + with mock.patch.dict(build.os.environ, {"ALLOW_RELEASED_ONNX_OPSET_ONLY": "1"}, clear=True): + for use_cuda in (False, True): + with self.subTest(use_cuda=use_cuda): + self.assertEqual( + build.get_onnx_backend_test_environment(use_cuda), + _UNRELEASED_OPSET_ENVIRONMENT, + ) + + if __name__ == "__main__": unittest.main() diff --git a/tools/python/materialize_onnx_node_tests.py b/tools/python/materialize_onnx_node_tests.py index 5071dfca75a70..0e4b6de46c924 100644 --- a/tools/python/materialize_onnx_node_tests.py +++ b/tools/python/materialize_onnx_node_tests.py @@ -28,8 +28,7 @@ * ``onnx.__version__ == --expected-onnx-version`` -- HARD FAIL. The model opset/IR is stamped from the *compiled* schema registry (``get_schema().since_version``), so a mismatched wheel silently bakes the - wrong opset into ``model.onnx``. Compared on the release base so a pre-release - (``rcN`` / ``.dev``) wheel of the pinned tag does not false-FATAL. + wrong opset into ``model.onnx``. The source and wheel pins must match exactly. * ``numpy.__version__`` vs ``--expected-numpy-version`` -- SOFT WARNING only, never a hard gate. ONNX's own corpus-gen numpy is uncontrollable/unknowable, so a fixed pin cannot guarantee byte-identity and a FATAL would false-red on @@ -52,7 +51,6 @@ import argparse import os -import re import shutil import sys import warnings @@ -93,19 +91,6 @@ class MaterializeError(RuntimeError): # --------------------------------------------------------------------------- # # Version / count guards # --------------------------------------------------------------------------- # -def _release_base(version: str) -> str: - """Return the release base (major.minor.micro) of a possibly-prerelease version. - - The ONNX opset-bump workflow ships release-candidate / dev wheels like - ``1.23.0rc1`` or ``1.23.0.dev20240101`` whose COMPILED opset registry already - matches the formal ``1.23.0`` tag. Comparing on the release base lets an RC - wheel of the pinned tag pass, while a genuine major/minor/micro mismatch - (e.g. ``1.22.0`` vs ``1.23.0``) still fails loud. - """ - m = re.match(r"^\s*(\d+)\.(\d+)\.(\d+)", version) - return ".".join(m.groups()) if m else version.strip() - - def _check_versions(expected_onnx: str, expected_numpy: str | None) -> None: """Assert wheel parity BEFORE importing/collecting cases. @@ -124,17 +109,14 @@ def _check_versions(expected_onnx: str, expected_numpy: str | None) -> None: import numpy # noqa: PLC0415 import onnx # noqa: PLC0415 - # onnx: HARD FAIL, RC/pre-release aware (compare release base, not the raw - # string) so an rcN/dev wheel of the pinned tag isn't a false FATAL. - if expected_onnx and _release_base(onnx.__version__) != _release_base(expected_onnx): + # onnx: HARD FAIL on any source/wheel version mismatch. + if expected_onnx and onnx.__version__ != expected_onnx: raise MaterializeError( f"onnx version mismatch: installed onnx=={onnx.__version__} but " - f"--expected-onnx-version=={expected_onnx} (release base " - f"{_release_base(onnx.__version__)} != {_release_base(expected_onnx)}). " + f"--expected-onnx-version=={expected_onnx}. " f"The node-test corpus opset/IR is baked from the COMPILED onnx schema " f"registry, so a mismatched wheel produces a silently-drifted corpus. " - f"Install onnx=={expected_onnx} (the cmake/deps.txt pin) before materializing. " - f"A pre-release (rcN/.dev) of the SAME release base is accepted." + f"Install onnx=={expected_onnx} (the cmake/deps.txt pin) before materializing." ) # numpy: SOFT WARNING, never a hard gate. ONNX's OWN corpus-gen numpy is # uncontrollable and unknowable, so a fixed pin can neither guarantee @@ -197,11 +179,9 @@ def _write_case(case, output_root: str) -> None: n_graph_in = len(case.model.graph.input) n_graph_out = len(case.model.graph.output) for i, (inputs, outputs) in enumerate(case.data_sets): - # Fail loud BEFORE writing any .pb for this dataset: the writer indexes - # graph.input[j]/graph.output[j] per collected value, so more collected - # values than graph arity would raise a bare IndexError mid-write (not - # caught by main's MaterializeError handler) and leave a partial tree. - if len(inputs) > n_graph_in or len(outputs) > n_graph_out: + # Fail loud before writing a partial dataset if positional values do not + # exactly match their graph-value declarations. + if len(inputs) != n_graph_in or len(outputs) != n_graph_out: raise MaterializeError( f"node case {case.name!r} data_set {i}: arity mismatch -- collected " f"{len(inputs)} input(s)/{len(outputs)} output(s) but the model graph "