From 333828948b9c80ad49c3d62d71f43cc07228a8a8 Mon Sep 17 00:00:00 2001 From: Chris Austen Date: Fri, 29 Sep 2023 10:25:04 -0400 Subject: [PATCH 01/19] Enable MLIR to be built with MIGraphX (#2184) Enable MLIR performance enhancements with MIGRAPHX_ENABLE_MLIR=1 --- .github/workflows/ci.yaml | 1 - Dockerfile | 4 ---- hip-clang.docker | 3 --- mlir-requirements.txt | 24 ------------------------ requirements.txt | 1 + src/targets/gpu/CMakeLists.txt | 2 +- 6 files changed, 2 insertions(+), 33 deletions(-) delete mode 100644 mlir-requirements.txt diff --git a/.github/workflows/ci.yaml b/.github/workflows/ci.yaml index f01d44c7c0d..aad99c2cd1f 100644 --- a/.github/workflows/ci.yaml +++ b/.github/workflows/ci.yaml @@ -137,7 +137,6 @@ jobs: -DMIGRAPHX_ENABLE_GPU=On \ -DMIGRAPHX_ENABLE_CPU=On \ -DMIGRAPHX_ENABLE_FPGA=On \ - -DMIGRAPHX_ENABLE_MLIR=On \ -DBUILD_DEV=On \ -DROCM_ENABLE_GH_ANNOTATIONS=On \ -DCLANG_TIDY_DEPEND_ON_TARGET=Off \ diff --git a/Dockerfile b/Dockerfile index 0e6416ab1ac..fcae5a21959 100644 --- a/Dockerfile +++ b/Dockerfile @@ -101,10 +101,6 @@ RUN cget -p $PREFIX install facebook/zstd@v1.4.5 -X subdir -DCMAKE_DIR=build/cma RUN cget -p $PREFIX install ccache@v4.1 -DENABLE_TESTING=OFF RUN cget -p /opt/cmake install kitware/cmake@v3.26.4 -# Install MLIR -ADD mlir-requirements.txt /mlir-requirements.txt -RUN cget -p /usr/local install -f /mlir-requirements.txt - COPY ./test/onnx/.onnxrt-commit / ARG ONNXRUNTIME_REPO=https://github.com/Microsoft/onnxruntime diff --git a/hip-clang.docker b/hip-clang.docker index e7d0c63becc..b1c805f2044 100755 --- a/hip-clang.docker +++ b/hip-clang.docker @@ -60,6 +60,3 @@ RUN pip3 install cmake==3.22.1 COPY ./tools/install_prereqs.sh / RUN /install_prereqs.sh /usr/local / && rm /install_prereqs.sh -# Install MLIR -ADD mlir-requirements.txt /mlir-requirements.txt -RUN cget -p /usr/local install -f /mlir-requirements.txt diff --git a/mlir-requirements.txt b/mlir-requirements.txt deleted file mode 100644 index 02af8d38df6..00000000000 --- a/mlir-requirements.txt +++ /dev/null @@ -1,24 +0,0 @@ -##################################################################################### -# The MIT License (MIT) -# -# Copyright (c) 2015-2022 Advanced Micro Devices, Inc. All rights reserved. -# -# Permission is hereby granted, free of charge, to any person obtaining a copy -# of this software and associated documentation files (the "Software"), to deal -# in the Software without restriction, including without limitation the rights -# to use, copy, modify, merge, publish, distribute, sublicense, and/or sell -# copies of the Software, and to permit persons to whom the Software is -# furnished to do so, subject to the following conditions: -# -# The above copyright notice and this permission notice shall be included in -# all copies or substantial portions of the Software. -# -# THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR -# IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, -# FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE -# AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER -# LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, -# OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN -# THE SOFTWARE. -##################################################################################### -ROCmSoftwarePlatform/rocMLIR@2c519c48eaa278d13e6c40bc0941119826d71512 -DBUILD_FAT_LIBROCKCOMPILER=On diff --git a/requirements.txt b/requirements.txt index 623fb480ff7..fcb5dcb13e5 100755 --- a/requirements.txt +++ b/requirements.txt @@ -29,3 +29,4 @@ pybind/pybind11@d159a563383d10c821ba7b2a71905d1207db6de4 --build msgpack/msgpack-c@cpp-3.3.0 -DMSGPACK_BUILD_TESTS=Off sqlite3@3.17 -DCMAKE_POSITION_INDEPENDENT_CODE=On ROCmSoftwarePlatform/composable_kernel@5172ec5280f14974beee2acf1af1db3b2670244c -DCK_BUILD_JIT_LIB=On -DCMAKE_POSITION_INDEPENDENT_CODE=On +ROCmSoftwarePlatform/rocMLIR@a48dfb1f163fb0b38369e73e580968b72e85b594 -DBUILD_FAT_LIBROCKCOMPILER=On diff --git a/src/targets/gpu/CMakeLists.txt b/src/targets/gpu/CMakeLists.txt index 9bbc4b61094..d6dcd182f8e 100644 --- a/src/targets/gpu/CMakeLists.txt +++ b/src/targets/gpu/CMakeLists.txt @@ -198,7 +198,7 @@ register_op(migraphx_gpu HEADER migraphx/gpu/convolution.hpp rocm_set_soversion(migraphx_gpu ${MIGRAPHX_SO_VERSION}) rocm_clang_tidy_check(migraphx_gpu) -set(MIGRAPHX_ENABLE_MLIR OFF CACHE BOOL "") +set(MIGRAPHX_ENABLE_MLIR ON CACHE BOOL "") if(MIGRAPHX_ENABLE_MLIR) # Find package rocMLIR From 4188c38e4f33dbf21a8085306eeeb596e24dd8d6 Mon Sep 17 00:00:00 2001 From: Umang Yadav <29876643+umangyadav@users.noreply.github.com> Date: Fri, 29 Sep 2023 19:33:06 -0400 Subject: [PATCH 02/19] Changes for the CK + HIPRTC (#2251) add flags for ck, Enable CK with hipRTC. CK can be used with the MIGRAPHX_ENABLE_CK=1 and MIGRAPHX_TUNE_CK=1 --- Jenkinsfile | 6 +++--- requirements.txt | 2 +- src/targets/gpu/compile_hip.cpp | 4 ++-- 3 files changed, 6 insertions(+), 6 deletions(-) diff --git a/Jenkinsfile b/Jenkinsfile index 49f30cf34cc..147363d6589 100755 --- a/Jenkinsfile +++ b/Jenkinsfile @@ -132,10 +132,10 @@ rocmtest clang_debug: rocmnode('cdna') { cmake_build -> cmake_build(flags: "-DCMAKE_BUILD_TYPE=debug -DMIGRAPHX_ENABLE_PYTHON=Off -DMIGRAPHX_ENABLE_MLIR=On -DCMAKE_CXX_FLAGS_DEBUG='${debug_flags_cxx}' -DCMAKE_C_FLAGS_DEBUG='${debug_flags}' -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')") } } -}, ck_release: rocmnode('mi100+') { cmake_build -> - stage('CK Release') { +}, ck_hiprtc: rocmnode('mi100+') { cmake_build -> + stage('CK hipRTC') { withEnv(['MIGRAPHX_ENABLE_CK=1', 'MIGRAPHX_TUNE_CK=1']) { - cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')") + cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DMIGRAPHX_USE_HIPRTC=On -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')") } } }, clang_asan: rocmnode('nogpu') { cmake_build -> diff --git a/requirements.txt b/requirements.txt index fcb5dcb13e5..6e24e0110a4 100755 --- a/requirements.txt +++ b/requirements.txt @@ -28,5 +28,5 @@ ROCmSoftwarePlatform/half@rocm-5.6.0 pybind/pybind11@d159a563383d10c821ba7b2a71905d1207db6de4 --build msgpack/msgpack-c@cpp-3.3.0 -DMSGPACK_BUILD_TESTS=Off sqlite3@3.17 -DCMAKE_POSITION_INDEPENDENT_CODE=On -ROCmSoftwarePlatform/composable_kernel@5172ec5280f14974beee2acf1af1db3b2670244c -DCK_BUILD_JIT_LIB=On -DCMAKE_POSITION_INDEPENDENT_CODE=On +ROCmSoftwarePlatform/composable_kernel@a22e479b8e1557961039db2d5c5ff89cff35e86b -DCK_BUILD_JIT_LIB=On -DCMAKE_POSITION_INDEPENDENT_CODE=On ROCmSoftwarePlatform/rocMLIR@a48dfb1f163fb0b38369e73e580968b72e85b594 -DBUILD_FAT_LIBROCKCOMPILER=On diff --git a/src/targets/gpu/compile_hip.cpp b/src/targets/gpu/compile_hip.cpp index a4afa5310ae..30565fadfe6 100644 --- a/src/targets/gpu/compile_hip.cpp +++ b/src/targets/gpu/compile_hip.cpp @@ -28,6 +28,7 @@ #include #include #include +#include #ifdef MIGRAPHX_USE_HIPRTC #include @@ -92,7 +93,7 @@ struct hiprtc_program { struct string_array { - std::vector strings{}; + std::deque strings{}; std::vector c_strs{}; string_array() {} @@ -209,7 +210,6 @@ std::vector> compile_hip_src_with_hiprtc(std::vector Date: Sun, 1 Oct 2023 15:07:34 -0400 Subject: [PATCH 03/19] Fix Jenkinsfile failure (#2265) --- Jenkinsfile | 22 ++++++++++++++++------ 1 file changed, 16 insertions(+), 6 deletions(-) diff --git a/Jenkinsfile b/Jenkinsfile index 147363d6589..bad38fd18e9 100755 --- a/Jenkinsfile +++ b/Jenkinsfile @@ -1,3 +1,7 @@ +def getgputargets() { + targets="gfx908;gfx90a;gfx1030;gfx1100;gfx1101;gfx1102" + return targets +} // def rocmtestnode(variant, name, body, args, pre) { def rocmtestnode(Map conf) { @@ -107,11 +111,13 @@ rocmtest clang_debug: rocmnode('cdna') { cmake_build -> stage('hipRTC Debug') { def sanitizers = "undefined" def debug_flags = "-g -O2 -fsanitize=${sanitizers} -fno-sanitize-recover=${sanitizers}" - cmake_build(flags: "-DCMAKE_BUILD_TYPE=debug -DMIGRAPHX_ENABLE_PYTHON=Off -DCMAKE_CXX_FLAGS_DEBUG='${debug_flags}' -DCMAKE_C_FLAGS_DEBUG='${debug_flags}' -DMIGRAPHX_USE_HIPRTC=On -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')", gpu_debug: true) + def gpu_targets = getgputargets() + cmake_build(flags: "-DCMAKE_BUILD_TYPE=debug -DMIGRAPHX_ENABLE_PYTHON=Off -DCMAKE_CXX_FLAGS_DEBUG='${debug_flags}' -DCMAKE_C_FLAGS_DEBUG='${debug_flags}' -DMIGRAPHX_USE_HIPRTC=On -DGPU_TARGETS='${gpu_targets}'", gpu_debug: true) } }, clang_release: rocmnode('mi100+') { cmake_build -> stage('Hip Clang Release') { - cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')") + def gpu_targets = getgputargets() + cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DGPU_TARGETS='${gpu_targets}'") stash includes: 'build/*.deb', name: 'migraphx-package' } // }, hidden_symbols: rocmnode('cdna') { cmake_build -> @@ -120,7 +126,8 @@ rocmtest clang_debug: rocmnode('cdna') { cmake_build -> // } }, all_targets_debug : rocmnode('cdna') { cmake_build -> stage('All targets Release') { - cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DMIGRAPHX_ENABLE_GPU=On -DMIGRAPHX_ENABLE_CPU=On -DMIGRAPHX_ENABLE_FPGA=On -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')") + def gpu_targets = getgputargets() + cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DMIGRAPHX_ENABLE_GPU=On -DMIGRAPHX_ENABLE_CPU=On -DMIGRAPHX_ENABLE_FPGA=On -DGPU_TARGETS='${gpu_targets}'") } }, mlir_debug: rocmnode('cdna') { cmake_build -> stage('MLIR Debug') { @@ -129,20 +136,23 @@ rocmtest clang_debug: rocmnode('cdna') { cmake_build -> // Note: the -fno-sanitize= is copied from upstream LLVM_UBSAN_FLAGS. def debug_flags_cxx = "-g -O2 -fsanitize=${sanitizers} -fno-sanitize=vptr,function -fno-sanitize-recover=${sanitizers}" def debug_flags = "-g -O2 -fsanitize=${sanitizers} -fno-sanitize=vptr -fno-sanitize-recover=${sanitizers}" - cmake_build(flags: "-DCMAKE_BUILD_TYPE=debug -DMIGRAPHX_ENABLE_PYTHON=Off -DMIGRAPHX_ENABLE_MLIR=On -DCMAKE_CXX_FLAGS_DEBUG='${debug_flags_cxx}' -DCMAKE_C_FLAGS_DEBUG='${debug_flags}' -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')") + def gpu_targets = getgputargets() + cmake_build(flags: "-DCMAKE_BUILD_TYPE=debug -DMIGRAPHX_ENABLE_PYTHON=Off -DMIGRAPHX_ENABLE_MLIR=On -DCMAKE_CXX_FLAGS_DEBUG='${debug_flags_cxx}' -DCMAKE_C_FLAGS_DEBUG='${debug_flags}' -DGPU_TARGETS='${gpu_targets}'") } } }, ck_hiprtc: rocmnode('mi100+') { cmake_build -> stage('CK hipRTC') { withEnv(['MIGRAPHX_ENABLE_CK=1', 'MIGRAPHX_TUNE_CK=1']) { - cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DMIGRAPHX_USE_HIPRTC=On -DGPU_TARGETS=$(/opt/rocm/bin/rocminfo | grep -o -m1 'gfx.*')") + def gpu_targets = getgputargets() + cmake_build(flags: "-DCMAKE_BUILD_TYPE=release -DMIGRAPHX_USE_HIPRTC=On -DGPU_TARGETS='${gpu_targets}'") } } }, clang_asan: rocmnode('nogpu') { cmake_build -> stage('Clang ASAN') { def sanitizers = "undefined,address" def debug_flags = "-g -O2 -fno-omit-frame-pointer -fsanitize=${sanitizers} -fno-sanitize-recover=${sanitizers}" - cmake_build(flags: "-DCMAKE_BUILD_TYPE=debug -DMIGRAPHX_ENABLE_PYTHON=Off -DMIGRAPHX_ENABLE_GPU=Off -DMIGRAPHX_ENABLE_CPU=On -DCMAKE_CXX_FLAGS_DEBUG='${debug_flags}' -DCMAKE_C_FLAGS_DEBUG='${debug_flags}'") + def gpu_targets = getgputargets() + cmake_build(flags: "-DCMAKE_BUILD_TYPE=debug -DMIGRAPHX_ENABLE_PYTHON=Off -DMIGRAPHX_ENABLE_GPU=Off -DMIGRAPHX_ENABLE_CPU=On -DCMAKE_CXX_FLAGS_DEBUG='${debug_flags}' -DCMAKE_C_FLAGS_DEBUG='${debug_flags}' -DGPU_TARGETS='${gpu_targets}'") } }//, clang_release_navi: rocmnode('navi21') { cmake_build -> // stage('HIP Clang Release Navi') { From ae5cc13ec2756160eb2ec261e4b1fca4427306ef Mon Sep 17 00:00:00 2001 From: Chris Austen Date: Mon, 2 Oct 2023 12:26:35 -0400 Subject: [PATCH 04/19] Add gfx906 target to Jenkins (#2269) --- Jenkinsfile | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/Jenkinsfile b/Jenkinsfile index bad38fd18e9..bf254937d24 100755 --- a/Jenkinsfile +++ b/Jenkinsfile @@ -1,5 +1,5 @@ def getgputargets() { - targets="gfx908;gfx90a;gfx1030;gfx1100;gfx1101;gfx1102" + targets="gfx906;gfx908;gfx90a;gfx1030;gfx1100;gfx1101;gfx1102" return targets } From 36eaf9e532828fa014e7ed2eebb40911ed73324b Mon Sep 17 00:00:00 2001 From: Umang Yadav <29876643+umangyadav@users.noreply.github.com> Date: Mon, 2 Oct 2023 22:54:17 -0400 Subject: [PATCH 05/19] just use one flush call (#2272) --- .../gpu/device/include/migraphx/gpu/device/launch.hpp | 8 ++++++++ 1 file changed, 8 insertions(+) diff --git a/src/targets/gpu/device/include/migraphx/gpu/device/launch.hpp b/src/targets/gpu/device/include/migraphx/gpu/device/launch.hpp index fcc98c941a8..b36ef2b851f 100644 --- a/src/targets/gpu/device/include/migraphx/gpu/device/launch.hpp +++ b/src/targets/gpu/device/include/migraphx/gpu/device/launch.hpp @@ -81,6 +81,14 @@ inline auto launch(hipStream_t stream, index_int global, index_int local) using f_type = decltype(f); dim3 nblocks(global / local); dim3 nthreads(local); + /* + hipGetLastError() returns error for the first failed HIP call that happened previously. + MIGraphX calls into various backend libraries and failed HIP calls can also happen there. + Calling hipGetLastError() would reset error code to hipSuccess, so that inside MIGraphX + failed call to hipLaunchKernelGGL() can be captured. + */ + hipError_t flush_call = hipGetLastError(); + (void)(flush_call); // cppcheck-suppress UseDeviceLaunch hipLaunchKernelGGL((launcher), nblocks, nthreads, 0, stream, f); hipError_t kernel_launch_status = hipGetLastError(); From 63e35438c93595847a41af13a9f3dfe80e2f8a01 Mon Sep 17 00:00:00 2001 From: Charlie Lin Date: Mon, 2 Oct 2023 23:14:00 -0400 Subject: [PATCH 06/19] Dynamic ConstantOfShape (#2111) Makes ConstantOfShape work with a variable dimensions input --- src/onnx/parse_constant_of_shape.cpp | 47 ++++++----- test/onnx/const_of_shape_default_test.onnx | 10 +++ test/onnx/const_of_shape_dyn_float_test.onnx | Bin 0 -> 189 bytes test/onnx/const_of_shape_dyn_int64_test.onnx | 14 ++++ .../onnx/const_of_shape_empty_input_test.onnx | Bin 208 -> 208 bytes test/onnx/const_of_shape_float_test.onnx | Bin 188 -> 204 bytes test/onnx/const_of_shape_int64_test.onnx | 10 +-- .../const_of_shape_no_value_attr_test.onnx | 8 +- test/onnx/gen_onnx.py | 74 +++++++++++++++--- test/onnx/onnx_test.cpp | 60 +++++++++++++- 10 files changed, 182 insertions(+), 41 deletions(-) create mode 100644 test/onnx/const_of_shape_default_test.onnx create mode 100644 test/onnx/const_of_shape_dyn_float_test.onnx create mode 100644 test/onnx/const_of_shape_dyn_int64_test.onnx diff --git a/src/onnx/parse_constant_of_shape.cpp b/src/onnx/parse_constant_of_shape.cpp index d3dceec4561..75d2f4e9beb 100644 --- a/src/onnx/parse_constant_of_shape.cpp +++ b/src/onnx/parse_constant_of_shape.cpp @@ -49,6 +49,8 @@ struct parse_constant_of_shape : op_parser { MIGRAPHX_THROW("ConstantOfShape: attribute value can contain only 1 elements!"); } + // convert to a scalar literal + l_val = literal(shape{l_val.get_shape().type(), {1}, {0}}, l_val.data()); } else { @@ -64,30 +66,37 @@ struct parse_constant_of_shape : op_parser migraphx::shape s; // input is empty, output is a scalar auto type = l_val.get_shape().type(); - // empty input tensor, output is a scalar - if(args[0]->get_shape().elements() == 0) + migraphx::argument input = args[0]->eval(); + if(not input.empty()) { - s = migraphx::shape{type, {1}, {0}}; + // empty input tensor, output is a scalar + if(args[0]->get_shape().elements() == 0) + { + s = migraphx::shape{type, {1}, {0}}; + } + else + { + std::vector dims; + input.visit([&](auto ia) { dims.assign(ia.begin(), ia.end()); }); + s = migraphx::shape{type, dims}; + } + literal l_out{}; + l_val.visit([&](auto val) { + using val_type = std::remove_cv_t; + // l_val contains only one element + std::vector out_vec(s.elements(), val.front()); + l_out = literal(s, out_vec); + }); + return info.add_literal(l_out); } + // has variable input (dynamic shape buffer) else { - migraphx::argument in = args[0]->eval(); - check_arg_empty(in, "ConstantOfShape: dynamic shape is not supported"); - - std::vector dims; - in.visit([&](auto input) { dims.assign(input.begin(), input.end()); }); - s = migraphx::shape{type, dims}; + auto dv_lit = info.add_literal(l_val); + auto alloc_ins = + info.add_instruction(make_op("allocate", {{"buf_type", type}}), args[0]); + return info.add_instruction(make_op("fill"), dv_lit, alloc_ins); } - - literal l_out{}; - l_val.visit([&](auto val) { - using val_type = std::remove_cv_t; - // l_val contains only one element - std::vector out_vec(s.elements(), val.front()); - l_out = literal(s, out_vec); - }); - - return info.add_literal(l_out); } } }; diff --git a/test/onnx/const_of_shape_default_test.onnx b/test/onnx/const_of_shape_default_test.onnx new file mode 100644 index 00000000000..463ad36b5aa --- /dev/null +++ b/test/onnx/const_of_shape_default_test.onnx @@ -0,0 +1,10 @@ +const_of_shape_default_test:‹ +6shape"Constant*# +value*:B shape_tensor  + +shapey"ConstantOfShapeconst_of_shape_default_testb +y + + + +B \ No newline at end of file diff --git a/test/onnx/const_of_shape_dyn_float_test.onnx b/test/onnx/const_of_shape_dyn_float_test.onnx new file mode 100644 index 0000000000000000000000000000000000000000..08d34f1e0a0889d340c2d1abd23fc3b3df5e1980 GIT binary patch literal 189 zcmdc!80LgJU9-Dc!80LgMA{CDTMCqJadd_literal(migraphx::literal(output_dims_shape, {2, 3, 4})); + migraphx::shape output_shape{migraphx::shape::float_type, {2, 3, 4}}; + std::vector vec(output_shape.elements(), 0.0); + mm->add_literal(migraphx::literal(output_shape, vec)); + + auto prog = optimize_onnx("const_of_shape_default_test.onnx"); + EXPECT(p == prog); +} + TEST_CASE(const_of_shape_empty_input_test) { migraphx::program p; auto* mm = p.get_main_module(); - mm->add_literal(migraphx::literal(migraphx::shape::int32_type)); + mm->add_literal(migraphx::literal(migraphx::shape::int64_type)); migraphx::shape s(migraphx::shape::int64_type, {1}, {0}); std::vector vec(s.elements(), 10); mm->add_literal(migraphx::literal(s, vec)); @@ -1057,7 +1071,7 @@ TEST_CASE(const_of_shape_float_test) { migraphx::program p; auto* mm = p.get_main_module(); - migraphx::shape ss(migraphx::shape::int32_type, {3}); + migraphx::shape ss(migraphx::shape::int64_type, {3}); mm->add_literal(migraphx::literal(ss, {2, 3, 4})); migraphx::shape s(migraphx::shape::float_type, {2, 3, 4}); std::vector vec(s.elements(), 10.0f); @@ -1071,8 +1085,10 @@ TEST_CASE(const_of_shape_int64_test) { migraphx::program p; auto* mm = p.get_main_module(); - migraphx::shape ss(migraphx::shape::int32_type, {3}); + // output_dims + migraphx::shape ss(migraphx::shape::int64_type, {3}); mm->add_literal(migraphx::literal(ss, {2, 3, 4})); + // constant shape literal migraphx::shape s(migraphx::shape::int64_type, {2, 3, 4}); std::vector vec(s.elements(), 10); mm->add_literal(migraphx::literal(s, vec)); @@ -1085,7 +1101,7 @@ TEST_CASE(const_of_shape_no_value_attr_test) { migraphx::program p; auto* mm = p.get_main_module(); - migraphx::shape ss(migraphx::shape::int32_type, {3}); + migraphx::shape ss(migraphx::shape::int64_type, {3}); mm->add_literal(migraphx::literal(ss, {2, 3, 4})); migraphx::shape s(migraphx::shape::float_type, {2, 3, 4}); std::vector vec(s.elements(), 0.0f); @@ -1095,6 +1111,42 @@ TEST_CASE(const_of_shape_no_value_attr_test) EXPECT(p == prog); } +TEST_CASE(const_of_shape_dyn_float_test) +{ + migraphx::program p; + auto* mm = p.get_main_module(); + auto od_param = + mm->add_parameter("output_dims", migraphx::shape{migraphx::shape::int64_type, {3}}); + auto alloc_ins = mm->add_instruction( + migraphx::make_op("allocate", {{"buf_type", migraphx::shape::float_type}}), od_param); + migraphx::shape dv_shape(migraphx::shape::float_type, {1}, {0}); + auto dv_lit = mm->add_literal(migraphx::literal(dv_shape, {10})); + auto fill_ins = mm->add_instruction(migraphx::make_op("fill"), dv_lit, alloc_ins); + mm->add_return({fill_ins}); + + migraphx::onnx_options options; + auto prog = parse_onnx("const_of_shape_dyn_float_test.onnx", options); + EXPECT(p == prog); +} + +TEST_CASE(const_of_shape_dyn_int64_test) +{ + migraphx::program p; + auto* mm = p.get_main_module(); + auto od_param = + mm->add_parameter("output_dims", migraphx::shape{migraphx::shape::int64_type, {3}}); + auto alloc_ins = mm->add_instruction( + migraphx::make_op("allocate", {{"buf_type", migraphx::shape::int64_type}}), od_param); + migraphx::shape dv_shape(migraphx::shape::int64_type, {1}, {0}); + auto dv_lit = mm->add_literal(migraphx::literal(dv_shape, {10})); + auto fill_ins = mm->add_instruction(migraphx::make_op("fill"), dv_lit, alloc_ins); + mm->add_return({fill_ins}); + + migraphx::onnx_options options; + auto prog = parse_onnx("const_of_shape_dyn_int64_test.onnx", options); + EXPECT(p == prog); +} + TEST_CASE(conv_autopad_fail_test) { EXPECT(test::throws([&] { optimize_onnx("conv_autopad_fail_test.onnx"); })); From 60b8b09718b7911f63d8238eff219ea5bb19d8c5 Mon Sep 17 00:00:00 2001 From: Attila Dusnoki <126579622+attila-dusnoki-htec@users.noreply.github.com> Date: Tue, 3 Oct 2023 15:23:45 +0200 Subject: [PATCH 07/19] Update ONNX python backend tests (#2179) --- test/py/onnx_backend_test.py | 1319 ++++++++++++++++++++++++++++------ 1 file changed, 1114 insertions(+), 205 deletions(-) diff --git a/test/py/onnx_backend_test.py b/test/py/onnx_backend_test.py index 31c4179da2b..30f8c53ab6e 100644 --- a/test/py/onnx_backend_test.py +++ b/test/py/onnx_backend_test.py @@ -64,44 +64,593 @@ def assert_similar_outputs(cls, ref_outputs, outputs, rtol, atol): def disabled_tests_onnx_1_7_0(backend_test): + # fails + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_argmax_keepdims_example_select_last_index_cpu') + backend_test.exclude( + r'test_argmax_negative_axis_keepdims_example_select_last_index_cpu') + backend_test.exclude( + r'test_argmax_no_keepdims_example_select_last_index_cpu') + backend_test.exclude(r'test_argmin_keepdims_example_select_last_index_cpu') + backend_test.exclude( + r'test_argmin_negative_axis_keepdims_example_select_last_index_cpu') + backend_test.exclude( + r'test_argmin_no_keepdims_example_select_last_index_cpu') backend_test.exclude(r'test_logsoftmax_axis_0_cpu') backend_test.exclude(r'test_logsoftmax_axis_1_cpu') backend_test.exclude(r'test_logsoftmax_default_axis_cpu') + backend_test.exclude(r'test_maxpool_2d_dilations_cpu') + backend_test.exclude(r'test_maxpool_with_argmax_2d_precomputed_pads_cpu') + backend_test.exclude( + r'test_maxpool_with_argmax_2d_precomputed_strides_cpu') + backend_test.exclude(r'test_nonmaxsuppression_center_point_box_format_cpu') + backend_test.exclude(r'test_nonmaxsuppression_flipped_coordinates_cpu') + backend_test.exclude(r'test_nonmaxsuppression_identical_boxes_cpu') + backend_test.exclude(r'test_nonmaxsuppression_limit_output_size_cpu') + backend_test.exclude( + r'test_nonmaxsuppression_suppress_by_IOU_and_scores_cpu') + backend_test.exclude(r'test_nonmaxsuppression_suppress_by_IOU_cpu') + backend_test.exclude(r'test_nonmaxsuppression_two_batches_cpu') + backend_test.exclude(r'test_nonmaxsuppression_two_classes_cpu') + backend_test.exclude(r'test_nonzero_example_cpu') + backend_test.exclude(r'test_round_cpu') backend_test.exclude(r'test_softmax_axis_0_cpu') backend_test.exclude(r'test_softmax_axis_1_cpu') backend_test.exclude(r'test_softmax_default_axis_cpu') + # from OnnxBackendPyTorchConvertedModelTest + backend_test.exclude(r'test_ConvTranspose2d_cpu') + backend_test.exclude(r'test_ConvTranspose2d_no_bias_cpu') -def disabled_tests_onnx_1_8_1(backend_test): - backend_test.exclude(r'test_if_seq_cpu') + # from OnnxBackendPyTorchOperatorModelTest + backend_test.exclude(r'test_operator_add_broadcast_cpu') + backend_test.exclude(r'test_operator_add_size1_right_broadcast_cpu') + backend_test.exclude(r'test_operator_addconstant_cpu') + backend_test.exclude(r'test_operator_convtranspose_cpu') + + # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_bitshift_left_uint16_cpu') + backend_test.exclude(r'test_bitshift_left_uint32_cpu') + backend_test.exclude(r'test_bitshift_left_uint64_cpu') + backend_test.exclude(r'test_bitshift_left_uint8_cpu') + backend_test.exclude(r'test_bitshift_right_uint16_cpu') + backend_test.exclude(r'test_bitshift_right_uint32_cpu') + backend_test.exclude(r'test_bitshift_right_uint64_cpu') + backend_test.exclude(r'test_bitshift_right_uint8_cpu') + backend_test.exclude(r'test_cast_FLOAT_to_STRING_cpu') + backend_test.exclude(r'test_cast_STRING_to_FLOAT_cpu') + backend_test.exclude(r'test_compress_0_cpu') + backend_test.exclude(r'test_compress_1_cpu') + backend_test.exclude(r'test_compress_default_axis_cpu') + backend_test.exclude(r'test_compress_negative_axis_cpu') + backend_test.exclude(r'test_constant_pad_cpu') + backend_test.exclude(r'test_convinteger_with_padding_cpu') + backend_test.exclude(r'test_convtranspose_1d_cpu') + backend_test.exclude(r'test_det_2d_cpu') + backend_test.exclude(r'test_det_nd_cpu') + backend_test.exclude(r'test_dynamicquantizelinear_cpu') + backend_test.exclude(r'test_dynamicquantizelinear_max_adjusted_cpu') + backend_test.exclude(r'test_dynamicquantizelinear_min_adjusted_cpu') + backend_test.exclude(r'test_edge_pad_cpu') + backend_test.exclude(r'test_einsum_batch_diagonal_cpu') + backend_test.exclude(r'test_einsum_batch_matmul_cpu') + backend_test.exclude(r'test_einsum_inner_prod_cpu') + backend_test.exclude(r'test_einsum_sum_cpu') + backend_test.exclude(r'test_einsum_transpose_cpu') + backend_test.exclude(r'test_hardmax_axis_0_cpu') + backend_test.exclude(r'test_hardmax_axis_1_cpu') + backend_test.exclude(r'test_hardmax_axis_2_cpu') + backend_test.exclude(r'test_hardmax_default_axis_cpu') + backend_test.exclude(r'test_hardmax_example_cpu') + backend_test.exclude(r'test_hardmax_negative_axis_cpu') + backend_test.exclude(r'test_hardmax_one_hot_cpu') + backend_test.exclude(r'test_isinf_cpu') + backend_test.exclude(r'test_isinf_negative_cpu') + backend_test.exclude(r'test_isinf_positive_cpu') + backend_test.exclude(r'test_matmulinteger_cpu') + backend_test.exclude(r'test_maxpool_2d_uint8_cpu') + backend_test.exclude(r'test_maxunpool_export_with_output_shape_cpu') + backend_test.exclude(r'test_maxunpool_export_without_output_shape_cpu') + backend_test.exclude(r'test_mod_mixed_sign_int32_cpu') + backend_test.exclude(r'test_mod_mixed_sign_int8_cpu') + backend_test.exclude(r'test_mvn_cpu') + backend_test.exclude( + r'test_negative_log_likelihood_loss_iinput_shape_is_NCd1_weight_ignore_index_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NC_cpu') + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1_cpu') + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1_ignore_index_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1_mean_weight_negative_ignore_index_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1_weight_cpu') + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_cpu') + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_no_weight_reduction_mean_ignore_index_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_reduction_mean_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_reduction_sum_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_with_weight_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_with_weight_reduction_mean_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_with_weight_reduction_sum_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2_with_weight_reduction_sum_ignore_index_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2d3_none_no_weight_negative_ignore_index_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2d3_sum_weight_high_ignore_index_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2d3d4d5_mean_weight_cpu' + ) + backend_test.exclude( + r'test_negative_log_likelihood_loss_input_shape_is_NCd1d2d3d4d5_none_no_weight_cpu' + ) + backend_test.exclude(r'test_qlinearconv_cpu') + backend_test.exclude(r'test_qlinearmatmul_2D_cpu') + backend_test.exclude(r'test_qlinearmatmul_3D_cpu') + backend_test.exclude(r'test_range_float_type_positive_delta_expanded_cpu') + backend_test.exclude(r'test_range_int32_type_negative_delta_expanded_cpu') + backend_test.exclude(r'test_reflect_pad_cpu') + backend_test.exclude( + r'test_resize_downsample_scales_cubic_A_n0p5_exclude_outside_cpu') + backend_test.exclude( + r'test_resize_downsample_scales_cubic_align_corners_cpu') + backend_test.exclude(r'test_resize_downsample_scales_cubic_cpu') + backend_test.exclude( + r'test_resize_downsample_scales_linear_align_corners_cpu') + backend_test.exclude(r'test_resize_downsample_scales_linear_cpu') + backend_test.exclude(r'test_resize_downsample_scales_nearest_cpu') + backend_test.exclude(r'test_resize_downsample_sizes_cubic_cpu') + backend_test.exclude( + r'test_resize_downsample_sizes_linear_pytorch_half_pixel_cpu') + backend_test.exclude(r'test_resize_downsample_sizes_nearest_cpu') + backend_test.exclude( + r'test_resize_downsample_sizes_nearest_tf_half_pixel_for_nn_cpu') + backend_test.exclude(r'test_resize_tf_crop_and_resize_cpu') + backend_test.exclude( + r'test_resize_upsample_scales_cubic_A_n0p5_exclude_outside_cpu') + backend_test.exclude( + r'test_resize_upsample_scales_cubic_align_corners_cpu') + backend_test.exclude(r'test_resize_upsample_scales_cubic_asymmetric_cpu') + backend_test.exclude(r'test_resize_upsample_scales_cubic_cpu') + backend_test.exclude( + r'test_resize_upsample_scales_linear_align_corners_cpu') + backend_test.exclude(r'test_resize_upsample_scales_linear_cpu') + backend_test.exclude(r'test_resize_upsample_scales_nearest_cpu') + backend_test.exclude(r'test_resize_upsample_sizes_cubic_cpu') + backend_test.exclude( + r'test_resize_upsample_sizes_nearest_ceil_half_pixel_cpu') + backend_test.exclude(r'test_resize_upsample_sizes_nearest_cpu') + backend_test.exclude( + r'test_resize_upsample_sizes_nearest_floor_align_corners_cpu') + backend_test.exclude( + r'test_resize_upsample_sizes_nearest_round_prefer_ceil_asymmetric_cpu') + backend_test.exclude(r'test_reversesequence_batch_cpu') + backend_test.exclude(r'test_reversesequence_time_cpu') + backend_test.exclude(r'test_scan9_sum_cpu') + backend_test.exclude(r'test_scan_sum_cpu') + backend_test.exclude(r'test_shrink_hard_cpu') + backend_test.exclude(r'test_shrink_soft_cpu') + backend_test.exclude(r'test_slice_cpu') + backend_test.exclude(r'test_slice_default_axes_cpu') + backend_test.exclude(r'test_slice_default_steps_cpu') + backend_test.exclude(r'test_slice_end_out_of_bounds_cpu') + backend_test.exclude(r'test_slice_neg_cpu') + backend_test.exclude(r'test_slice_neg_steps_cpu') + backend_test.exclude(r'test_slice_negative_axes_cpu') + backend_test.exclude(r'test_slice_start_out_of_bounds_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1_mean_weight_negative_ignore_index_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1_mean_weight_negative_ignore_index_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1_mean_weight_negative_ignore_index_log_prob_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1_mean_weight_negative_ignore_index_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_none_no_weight_negative_ignore_index_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_none_no_weight_negative_ignore_index_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_none_no_weight_negative_ignore_index_log_prob_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_none_no_weight_negative_ignore_index_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_sum_weight_high_ignore_index_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_sum_weight_high_ignore_index_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_sum_weight_high_ignore_index_log_prob_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3_sum_weight_high_ignore_index_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_mean_weight_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_mean_weight_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_mean_weight_log_prob_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_mean_weight_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_none_no_weight_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_none_no_weight_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_none_no_weight_log_prob_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_input_shape_is_NCd1d2d3d4d5_none_no_weight_log_prob_expanded_cpu' + ) + backend_test.exclude(r'test_softmax_cross_entropy_mean_3d_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_mean_3d_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_mean_3d_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_3d_log_prob_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_mean_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_mean_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_mean_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_log_prob_expanded_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_3d_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_3d_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_3d_log_prob_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_3d_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_4d_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_4d_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_4d_log_prob_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_4d_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_expanded_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_no_weight_ignore_index_log_prob_expanded_cpu' + ) + backend_test.exclude(r'test_softmax_cross_entropy_mean_weight_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_expanded_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_3d_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_3d_expanded_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_3d_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_3d_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_4d_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_4d_expanded_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_4d_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_4d_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_expanded_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_ignore_index_log_prob_expanded_cpu' + ) + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_mean_weight_log_prob_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_none_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_none_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_none_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_none_log_prob_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_none_weights_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_none_weights_expanded_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_none_weights_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_none_weights_log_prob_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_sum_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_sum_expanded_cpu') + backend_test.exclude(r'test_softmax_cross_entropy_sum_log_prob_cpu') + backend_test.exclude( + r'test_softmax_cross_entropy_sum_log_prob_expanded_cpu') + backend_test.exclude(r'test_split_zero_size_splits_cpu') + backend_test.exclude( + r'test_strnormalizer_export_monday_casesensintive_lower_cpu') + backend_test.exclude( + r'test_strnormalizer_export_monday_casesensintive_nochangecase_cpu') + backend_test.exclude( + r'test_strnormalizer_export_monday_casesensintive_upper_cpu') + backend_test.exclude(r'test_strnormalizer_export_monday_empty_output_cpu') + backend_test.exclude( + r'test_strnormalizer_export_monday_insensintive_upper_twodim_cpu') + backend_test.exclude(r'test_strnormalizer_nostopwords_nochangecase_cpu') + backend_test.exclude( + r'test_tfidfvectorizer_tf_batch_onlybigrams_skip0_cpu') + backend_test.exclude( + r'test_tfidfvectorizer_tf_batch_onlybigrams_skip5_cpu') + backend_test.exclude( + r'test_tfidfvectorizer_tf_batch_uniandbigrams_skip5_cpu') + backend_test.exclude(r'test_tfidfvectorizer_tf_only_bigrams_skip0_cpu') + backend_test.exclude(r'test_tfidfvectorizer_tf_onlybigrams_levelempty_cpu') + backend_test.exclude(r'test_tfidfvectorizer_tf_onlybigrams_skip5_cpu') + backend_test.exclude(r'test_tfidfvectorizer_tf_uniandbigrams_skip5_cpu') + backend_test.exclude(r'test_top_k_cpu') + backend_test.exclude(r'test_top_k_negative_axis_cpu') + backend_test.exclude(r'test_top_k_smallest_cpu') + backend_test.exclude(r'test_unique_not_sorted_without_axis_cpu') + backend_test.exclude(r'test_unique_sorted_with_axis_3d_cpu') + backend_test.exclude(r'test_unique_sorted_with_axis_cpu') + backend_test.exclude(r'test_unique_sorted_with_negative_axis_cpu') + backend_test.exclude(r'test_unique_sorted_without_axis_cpu') + backend_test.exclude(r'test_upsample_nearest_cpu') + + # from OnnxBackendPyTorchConvertedModelTest + backend_test.exclude(r'test_PReLU_1d_multiparam_cpu') + backend_test.exclude(r'test_PReLU_2d_multiparam_cpu') + backend_test.exclude(r'test_PReLU_3d_multiparam_cpu') + backend_test.exclude(r'test_ReplicationPad2d_cpu') + + # from OnnxBackendPyTorchOperatorModelTest + backend_test.exclude(r'test_operator_add_size1_broadcast_cpu') + backend_test.exclude(r'test_operator_add_size1_singleton_broadcast_cpu') + + # from OnnxBackendSimpleModelTest + backend_test.exclude(r'test_gradient_of_add_and_mul_cpu') + backend_test.exclude(r'test_gradient_of_add_cpu') + backend_test.exclude(r'test_sequence_model1_cpu') + backend_test.exclude(r'test_sequence_model2_cpu') + backend_test.exclude(r'test_sequence_model3_cpu') + backend_test.exclude(r'test_sequence_model4_cpu') + backend_test.exclude(r'test_sequence_model5_cpu') + backend_test.exclude(r'test_sequence_model6_cpu') + backend_test.exclude(r'test_sequence_model7_cpu') + backend_test.exclude(r'test_sequence_model8_cpu') + backend_test.exclude(r'test_shrink_cpu') + backend_test.exclude(r'test_strnorm_model_monday_casesensintive_lower_cpu') + backend_test.exclude( + r'test_strnorm_model_monday_casesensintive_nochangecase_cpu') + backend_test.exclude(r'test_strnorm_model_monday_casesensintive_upper_cpu') + backend_test.exclude(r'test_strnorm_model_monday_empty_output_cpu') + backend_test.exclude( + r'test_strnorm_model_monday_insensintive_upper_twodim_cpu') + backend_test.exclude(r'test_strnorm_model_nostopwords_nochangecase_cpu') + + +def disabled_tests_onnx_1_8_0(backend_test): + # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_cast_BFLOAT16_to_FLOAT_cpu') + backend_test.exclude(r'test_cast_FLOAT_to_BFLOAT16_cpu') backend_test.exclude(r'test_if_seq_cpu') - backend_test.exclude(r'test_reduce_sum_default_axes_keepdims_example_cpu') - backend_test.exclude(r'test_reduce_sum_default_axes_keepdims_random_cpu') - backend_test.exclude(r'test_reduce_sum_do_not_keepdims_example_cpu') - backend_test.exclude(r'test_reduce_sum_do_not_keepdims_random_cpu') - backend_test.exclude(r'test_reduce_sum_empty_axes_input_noop_example_cpu') + backend_test.exclude(r'test_loop11_cpu') + backend_test.exclude(r'test_loop13_seq_cpu') + backend_test.exclude(r'test_nllloss_NC_cpu') + backend_test.exclude(r'test_nllloss_NCd1_cpu') + backend_test.exclude(r'test_nllloss_NCd1_ii_cpu') + backend_test.exclude(r'test_nllloss_NCd1_mean_weight_negative_ii_cpu') + backend_test.exclude(r'test_nllloss_NCd1_weight_cpu') + backend_test.exclude(r'test_nllloss_NCd1_weight_ii_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2_cpu') + backend_test.exclude( + r'test_nllloss_NCd1d2_no_weight_reduction_mean_ii_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2_reduction_mean_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2_reduction_sum_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2_with_weight_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2_with_weight_reduction_mean_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2_with_weight_reduction_sum_cpu') + backend_test.exclude( + r'test_nllloss_NCd1d2_with_weight_reduction_sum_ii_cpu') + backend_test.exclude( + r'test_nllloss_NCd1d2d3_none_no_weight_negative_ii_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2d3_sum_weight_high_ii_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2d3d4d5_mean_weight_cpu') + backend_test.exclude(r'test_nllloss_NCd1d2d3d4d5_none_no_weight_cpu') backend_test.exclude(r'test_reduce_sum_empty_axes_input_noop_random_cpu') - backend_test.exclude(r'test_reduce_sum_keepdims_example_cpu') - backend_test.exclude(r'test_reduce_sum_keepdims_random_cpu') - backend_test.exclude(r'test_reduce_sum_negative_axes_keepdims_example_cpu') - backend_test.exclude(r'test_reduce_sum_negative_axes_keepdims_random_cpu') - backend_test.exclude(r'test_unsqueeze_axis_0_cpu') - backend_test.exclude(r'test_unsqueeze_axis_1_cpu') - backend_test.exclude(r'test_unsqueeze_axis_2_cpu') - backend_test.exclude(r'test_unsqueeze_negative_axes_cpu') - backend_test.exclude(r'test_unsqueeze_three_axes_cpu') - backend_test.exclude(r'test_unsqueeze_two_axes_cpu') - backend_test.exclude(r'test_unsqueeze_unsorted_axes_cpu') + backend_test.exclude(r'test_sce_NCd1_mean_weight_negative_ii_cpu') + backend_test.exclude(r'test_sce_NCd1_mean_weight_negative_ii_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1_mean_weight_negative_ii_log_prob_cpu') + backend_test.exclude( + r'test_sce_NCd1_mean_weight_negative_ii_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3_none_no_weight_negative_ii_cpu') + backend_test.exclude( + r'test_sce_NCd1d2d3_none_no_weight_negative_ii_expanded_cpu') + backend_test.exclude( + r'test_sce_NCd1d2d3_none_no_weight_negative_ii_log_prob_cpu') + backend_test.exclude( + r'test_sce_NCd1d2d3_none_no_weight_negative_ii_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3_sum_weight_high_ii_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3_sum_weight_high_ii_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3_sum_weight_high_ii_log_prob_cpu') + backend_test.exclude( + r'test_sce_NCd1d2d3_sum_weight_high_ii_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3d4d5_mean_weight_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3d4d5_mean_weight_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3d4d5_mean_weight_log_prob_cpu') + backend_test.exclude( + r'test_sce_NCd1d2d3d4d5_mean_weight_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3d4d5_none_no_weight_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3d4d5_none_no_weight_expanded_cpu') + backend_test.exclude(r'test_sce_NCd1d2d3d4d5_none_no_weight_log_prob_cpu') + backend_test.exclude( + r'test_sce_NCd1d2d3d4d5_none_no_weight_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_3d_cpu') + backend_test.exclude(r'test_sce_mean_3d_expanded_cpu') + backend_test.exclude(r'test_sce_mean_3d_log_prob_cpu') + backend_test.exclude(r'test_sce_mean_3d_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_cpu') + backend_test.exclude(r'test_sce_mean_expanded_cpu') + backend_test.exclude(r'test_sce_mean_log_prob_cpu') + backend_test.exclude(r'test_sce_mean_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_3d_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_3d_expanded_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_3d_log_prob_cpu') + backend_test.exclude( + r'test_sce_mean_no_weight_ii_3d_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_4d_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_4d_expanded_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_4d_log_prob_cpu') + backend_test.exclude( + r'test_sce_mean_no_weight_ii_4d_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_expanded_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_log_prob_cpu') + backend_test.exclude(r'test_sce_mean_no_weight_ii_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_cpu') + backend_test.exclude(r'test_sce_mean_weight_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_3d_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_3d_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_3d_log_prob_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_3d_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_4d_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_4d_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_4d_log_prob_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_4d_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_log_prob_cpu') + backend_test.exclude(r'test_sce_mean_weight_ii_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_mean_weight_log_prob_cpu') + backend_test.exclude(r'test_sce_mean_weight_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_none_cpu') + backend_test.exclude(r'test_sce_none_expanded_cpu') + backend_test.exclude(r'test_sce_none_log_prob_cpu') + backend_test.exclude(r'test_sce_none_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_none_weights_cpu') + backend_test.exclude(r'test_sce_none_weights_expanded_cpu') + backend_test.exclude(r'test_sce_none_weights_log_prob_cpu') + backend_test.exclude(r'test_sce_none_weights_log_prob_expanded_cpu') + backend_test.exclude(r'test_sce_sum_cpu') + backend_test.exclude(r'test_sce_sum_expanded_cpu') + backend_test.exclude(r'test_sce_sum_log_prob_cpu') + backend_test.exclude(r'test_sce_sum_log_prob_expanded_cpu') + backend_test.exclude(r'test_sequence_insert_at_back_cpu') + backend_test.exclude(r'test_sequence_insert_at_front_cpu') + backend_test.exclude(r'test_split_variable_parts_1d_cpu') + backend_test.exclude(r'test_split_variable_parts_2d_cpu') + backend_test.exclude(r'test_split_variable_parts_default_axis_cpu') + + +def disabled_tests_onnx_1_9_0(backend_test): + # fails + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_gru_batchwise_cpu') + backend_test.exclude(r'test_lstm_batchwise_cpu') + backend_test.exclude(r'test_simple_rnn_batchwise_cpu') + backend_test.exclude(r'test_tril_cpu') + backend_test.exclude(r'test_tril_one_row_neg_cpu') + backend_test.exclude(r'test_tril_square_cpu') + # from OnnxBackendPyTorchConvertedModelTest + backend_test.exclude(r'test_MaxPool1d_stride_padding_dilation_cpu') + backend_test.exclude(r'test_MaxPool2d_stride_padding_dilation_cpu') + + # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_convinteger_without_padding_cpu') + backend_test.exclude(r'test_convtranspose_autopad_same_cpu') + backend_test.exclude(r'test_identity_sequence_cpu') + backend_test.exclude(r'test_tril_neg_cpu') + backend_test.exclude(r'test_tril_out_neg_cpu') + backend_test.exclude(r'test_tril_out_pos_cpu') + backend_test.exclude(r'test_tril_pos_cpu') + backend_test.exclude(r'test_tril_square_neg_cpu') + backend_test.exclude(r'test_tril_zero_cpu') + backend_test.exclude(r'test_triu_neg_cpu') + backend_test.exclude(r'test_triu_one_row_cpu') + backend_test.exclude(r'test_triu_out_neg_out_cpu') + backend_test.exclude(r'test_triu_out_pos_cpu') + backend_test.exclude(r'test_triu_pos_cpu') + backend_test.exclude(r'test_triu_square_neg_cpu') + backend_test.exclude(r'test_triu_zero_cpu') def disabled_tests_onnx_1_10_0(backend_test): - # unsupported shape attributes - backend_test.exclude(r'test_shape_end_1_cpu') - backend_test.exclude(r'test_shape_end_negative_1_cpu') - backend_test.exclude(r'test_shape_start_1_cpu') - backend_test.exclude(r'test_shape_start_1_end_2_cpu') - backend_test.exclude(r'test_shape_start_1_end_negative_1_cpu') - backend_test.exclude(r'test_shape_start_negative_1_cpu') + # fails + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_bernoulli_double_expanded_cpu') + backend_test.exclude(r'test_bernoulli_expanded_cpu') + backend_test.exclude(r'test_bernoulli_seed_expanded_cpu') + + # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_bernoulli_cpu') + backend_test.exclude(r'test_bernoulli_double_cpu') + backend_test.exclude(r'test_bernoulli_seed_cpu') + backend_test.exclude(r'test_castlike_BFLOAT16_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_BFLOAT16_to_FLOAT_expanded_cpu') + backend_test.exclude(r'test_castlike_DOUBLE_to_FLOAT16_cpu') + backend_test.exclude(r'test_castlike_DOUBLE_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_FLOAT16_to_DOUBLE_cpu') + backend_test.exclude(r'test_castlike_FLOAT16_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_BFLOAT16_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_BFLOAT16_expanded_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_DOUBLE_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT16_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_STRING_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_STRING_expanded_cpu') + backend_test.exclude(r'test_castlike_STRING_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_STRING_to_FLOAT_expanded_cpu') + backend_test.exclude(r'test_optional_get_element_cpu') + backend_test.exclude(r'test_optional_get_element_sequence_cpu') + backend_test.exclude(r'test_optional_has_element_cpu') + backend_test.exclude(r'test_optional_has_element_empty_cpu') def disabled_tests_onnx_1_11_0(backend_test): @@ -109,22 +658,90 @@ def disabled_tests_onnx_1_11_0(backend_test): backend_test.exclude(r'test_scatter_elements_with_duplicate_indices_cpu') # fails + # from OnnxBackendNodeModelTest backend_test.exclude(r'test_roialign_aligned_false_cpu') backend_test.exclude(r'test_roialign_aligned_true_cpu') backend_test.exclude(r'test_scatternd_add_cpu') backend_test.exclude(r'test_scatternd_multiply_cpu') # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_gridsample_aligncorners_true_cpu') + backend_test.exclude(r'test_gridsample_bicubic_cpu') + backend_test.exclude(r'test_gridsample_bilinear_cpu') + backend_test.exclude(r'test_gridsample_border_padding_cpu') + backend_test.exclude(r'test_gridsample_cpu') + backend_test.exclude(r'test_gridsample_nearest_cpu') + backend_test.exclude(r'test_gridsample_reflection_padding_cpu') + backend_test.exclude(r'test_gridsample_zeros_padding_cpu') backend_test.exclude(r'test_identity_opt_cpu') backend_test.exclude(r'test_if_opt_cpu') + backend_test.exclude(r'test_loop16_seq_none_cpu') def disabled_tests_onnx_1_12_0(backend_test): - pass + # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_blackmanwindow_cpu') + backend_test.exclude(r'test_blackmanwindow_expanded_cpu') + backend_test.exclude(r'test_blackmanwindow_symmetric_cpu') + backend_test.exclude(r'test_blackmanwindow_symmetric_expanded_cpu') + backend_test.exclude(r'test_dft_axis_cpu') + backend_test.exclude(r'test_dft_cpu') + backend_test.exclude(r'test_dft_inverse_cpu') + backend_test.exclude(r'test_hammingwindow_cpu') + backend_test.exclude(r'test_hammingwindow_expanded_cpu') + backend_test.exclude(r'test_hammingwindow_symmetric_cpu') + backend_test.exclude(r'test_hammingwindow_symmetric_expanded_cpu') + backend_test.exclude(r'test_hannwindow_cpu') + backend_test.exclude(r'test_hannwindow_expanded_cpu') + backend_test.exclude(r'test_hannwindow_symmetric_cpu') + backend_test.exclude(r'test_hannwindow_symmetric_expanded_cpu') + backend_test.exclude(r'test_layer_normalization_2d_axis0_cpu') + backend_test.exclude(r'test_layer_normalization_2d_axis1_cpu') + backend_test.exclude(r'test_layer_normalization_2d_axis_negative_1_cpu') + backend_test.exclude(r'test_layer_normalization_2d_axis_negative_2_cpu') + backend_test.exclude(r'test_layer_normalization_3d_axis0_epsilon_cpu') + backend_test.exclude(r'test_layer_normalization_3d_axis1_epsilon_cpu') + backend_test.exclude(r'test_layer_normalization_3d_axis2_epsilon_cpu') + backend_test.exclude( + r'test_layer_normalization_3d_axis_negative_1_epsilon_cpu') + backend_test.exclude( + r'test_layer_normalization_3d_axis_negative_2_epsilon_cpu') + backend_test.exclude( + r'test_layer_normalization_3d_axis_negative_3_epsilon_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis0_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis1_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis2_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis3_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis_negative_1_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis_negative_2_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis_negative_3_cpu') + backend_test.exclude(r'test_layer_normalization_4d_axis_negative_4_cpu') + backend_test.exclude(r'test_layer_normalization_default_axis_cpu') + backend_test.exclude(r'test_melweightmatrix_cpu') + backend_test.exclude(r'test_sequence_map_add_1_sequence_1_tensor_cpu') + backend_test.exclude( + r'test_sequence_map_add_1_sequence_1_tensor_expanded_cpu') + backend_test.exclude(r'test_sequence_map_add_2_sequences_cpu') + backend_test.exclude(r'test_sequence_map_add_2_sequences_expanded_cpu') + backend_test.exclude(r'test_sequence_map_extract_shapes_cpu') + backend_test.exclude(r'test_sequence_map_extract_shapes_expanded_cpu') + backend_test.exclude(r'test_sequence_map_identity_1_sequence_1_tensor_cpu') + backend_test.exclude( + r'test_sequence_map_identity_1_sequence_1_tensor_expanded_cpu') + backend_test.exclude(r'test_sequence_map_identity_1_sequence_cpu') + backend_test.exclude(r'test_sequence_map_identity_1_sequence_expanded_cpu') + backend_test.exclude(r'test_sequence_map_identity_2_sequences_cpu') + backend_test.exclude( + r'test_sequence_map_identity_2_sequences_expanded_cpu') + backend_test.exclude(r'test_stft_cpu') + backend_test.exclude(r'test_stft_with_window_cpu') def disabled_tests_onnx_1_13_0(backend_test): # fails + # from OnnxBackendNodeModelTest backend_test.exclude(r'test_reduce_l1_do_not_keepdims_example_cpu') backend_test.exclude(r'test_reduce_l1_do_not_keepdims_random_cpu') backend_test.exclude(r'test_reduce_l1_keep_dims_example_cpu') @@ -158,11 +775,273 @@ def disabled_tests_onnx_1_13_0(backend_test): backend_test.exclude(r'test_scatternd_min_cpu') # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_bitwise_and_i16_3d_cpu') + backend_test.exclude(r'test_bitwise_and_i32_2d_cpu') + backend_test.exclude(r'test_bitwise_and_ui64_bcast_3v1d_cpu') + backend_test.exclude(r'test_bitwise_and_ui8_bcast_4v3d_cpu') + backend_test.exclude(r'test_bitwise_not_2d_cpu') + backend_test.exclude(r'test_bitwise_not_3d_cpu') + backend_test.exclude(r'test_bitwise_not_4d_cpu') + backend_test.exclude(r'test_bitwise_or_i16_4d_cpu') + backend_test.exclude(r'test_bitwise_or_i32_2d_cpu') + backend_test.exclude(r'test_bitwise_or_ui64_bcast_3v1d_cpu') + backend_test.exclude(r'test_bitwise_or_ui8_bcast_4v3d_cpu') + backend_test.exclude(r'test_bitwise_xor_i16_3d_cpu') + backend_test.exclude(r'test_bitwise_xor_i32_2d_cpu') + backend_test.exclude(r'test_bitwise_xor_ui64_bcast_3v1d_cpu') + backend_test.exclude(r'test_bitwise_xor_ui8_bcast_4v3d_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_and_pad_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_and_pad_expanded_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_axes_chw_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_axes_chw_expanded_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_axes_hwc_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_axes_hwc_expanded_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_expanded_cpu') + backend_test.exclude(r'test_center_crop_pad_pad_cpu') + backend_test.exclude(r'test_center_crop_pad_pad_expanded_cpu') + backend_test.exclude(r'test_col2im_5d_cpu') + backend_test.exclude(r'test_col2im_cpu') + backend_test.exclude(r'test_col2im_dilations_cpu') + backend_test.exclude(r'test_col2im_pads_cpu') + backend_test.exclude(r'test_col2im_strides_cpu') backend_test.exclude(r'test_constant_pad_axes_cpu') + backend_test.exclude(r'test_group_normalization_epsilon_cpu') + backend_test.exclude(r'test_group_normalization_epsilon_expanded_cpu') + backend_test.exclude(r'test_group_normalization_example_cpu') + backend_test.exclude(r'test_group_normalization_example_expanded_cpu') + backend_test.exclude(r'test_mish_cpu') + backend_test.exclude(r'test_mvn_expanded_ver18_cpu') + backend_test.exclude(r'test_optional_get_element_optional_sequence_cpu') + backend_test.exclude(r'test_optional_get_element_optional_tensor_cpu') + backend_test.exclude(r'test_optional_get_element_tensor_cpu') + backend_test.exclude( + r'test_optional_has_element_empty_no_input_name_optional_input_cpu') + backend_test.exclude( + r'test_optional_has_element_empty_no_input_name_tensor_input_cpu') + backend_test.exclude( + r'test_optional_has_element_empty_no_input_optional_input_cpu') + backend_test.exclude( + r'test_optional_has_element_empty_no_input_tensor_input_cpu') + backend_test.exclude(r'test_optional_has_element_empty_optional_input_cpu') + backend_test.exclude(r'test_optional_has_element_optional_input_cpu') + backend_test.exclude(r'test_optional_has_element_tensor_input_cpu') + backend_test.exclude(r'test_prelu_broadcast_expanded_cpu') + backend_test.exclude(r'test_prelu_example_expanded_cpu') backend_test.exclude(r'test_reduce_l1_default_axes_keepdims_example_cpu') + backend_test.exclude(r'test_reduce_l1_default_axes_keepdims_random_cpu') + backend_test.exclude(r'test_reduce_l2_default_axes_keepdims_example_cpu') + backend_test.exclude(r'test_reduce_l2_default_axes_keepdims_random_cpu') + backend_test.exclude( + r'test_reduce_log_sum_exp_default_axes_keepdims_example_cpu') + backend_test.exclude( + r'test_reduce_log_sum_exp_default_axes_keepdims_random_cpu') + backend_test.exclude( + r'test_reduce_sum_square_default_axes_keepdims_example_cpu') + backend_test.exclude( + r'test_reduce_sum_square_default_axes_keepdims_random_cpu') + backend_test.exclude(r'test_resize_downsample_scales_cubic_antialias_cpu') + backend_test.exclude(r'test_resize_downsample_scales_linear_antialias_cpu') + backend_test.exclude(r'test_resize_downsample_sizes_cubic_antialias_cpu') + backend_test.exclude(r'test_resize_downsample_sizes_linear_antialias_cpu') + backend_test.exclude( + r'test_resize_downsample_sizes_nearest_not_larger_cpu') + backend_test.exclude( + r'test_resize_downsample_sizes_nearest_not_smaller_cpu') + backend_test.exclude(r'test_resize_tf_crop_and_resize_axes_2_3_cpu') + backend_test.exclude(r'test_resize_tf_crop_and_resize_axes_3_2_cpu') + backend_test.exclude(r'test_resize_upsample_scales_nearest_axes_2_3_cpu') + backend_test.exclude(r'test_resize_upsample_scales_nearest_axes_3_2_cpu') + backend_test.exclude(r'test_resize_upsample_sizes_nearest_axes_2_3_cpu') + backend_test.exclude(r'test_resize_upsample_sizes_nearest_axes_3_2_cpu') + backend_test.exclude(r'test_resize_upsample_sizes_nearest_not_larger_cpu') + backend_test.exclude(r'test_scatter_elements_with_reduction_max_cpu') + backend_test.exclude(r'test_scatter_elements_with_reduction_min_cpu') + + # The following tests fail due to the CastLike operator being unsupported + backend_test.exclude(r'test_elu_default_expanded_ver18_cpu') + backend_test.exclude(r'test_elu_example_expanded_ver18_cpu') + backend_test.exclude(r'test_elu_expanded_ver18_cpu') + backend_test.exclude(r'test_hardsigmoid_default_expanded_ver18_cpu') + backend_test.exclude(r'test_hardsigmoid_example_expanded_ver18_cpu') + backend_test.exclude(r'test_hardsigmoid_expanded_ver18_cpu') + backend_test.exclude(r'test_leakyrelu_default_expanded_cpu') + backend_test.exclude(r'test_leakyrelu_example_expanded_cpu') + backend_test.exclude(r'test_leakyrelu_expanded_cpu') + backend_test.exclude(r'test_selu_default_expanded_ver18_cpu') + backend_test.exclude(r'test_selu_example_expanded_ver18_cpu') + backend_test.exclude(r'test_selu_expanded_ver18_cpu') + backend_test.exclude(r'test_shrink_hard_expanded_ver18_cpu') + backend_test.exclude(r'test_shrink_soft_expanded_ver18_cpu') + backend_test.exclude(r'test_split_1d_uneven_split_opset18_cpu') + backend_test.exclude(r'test_split_2d_uneven_split_opset18_cpu') + backend_test.exclude(r'test_thresholdedrelu_default_expanded_ver18_cpu') + backend_test.exclude(r'test_thresholdedrelu_example_expanded_ver18_cpu') + backend_test.exclude(r'test_thresholdedrelu_expanded_ver18_cpu') + backend_test.exclude(r'test_relu_expanded_ver18_cpu') + backend_test.exclude(r'test_softsign_example_expanded_ver18_cpu') + backend_test.exclude(r'test_softsign_expanded_ver18_cpu') + + +def disabled_tests_onnx_1_14_0(backend_test): + # fails + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_averagepool_2d_dilations_cpu') + backend_test.exclude(r'test_roialign_mode_max_cpu') + + # errors + # from OnnxBackendNodeModelTest + backend_test.exclude(r'test_basic_deform_conv_with_padding_cpu') + backend_test.exclude(r'test_basic_deform_conv_without_padding_cpu') + backend_test.exclude(r'test_center_crop_pad_crop_negative_axes_hwc_cpu') + backend_test.exclude( + r'test_center_crop_pad_crop_negative_axes_hwc_expanded_cpu') + backend_test.exclude(r'test_constant_pad_negative_axes_cpu') + backend_test.exclude(r'test_deform_conv_with_mask_bias_cpu') + backend_test.exclude(r'test_deform_conv_with_multiple_offset_groups_cpu') + backend_test.exclude(r'test_equal_string_broadcast_cpu') + backend_test.exclude(r'test_equal_string_cpu') + backend_test.exclude(r'test_lppool_1d_default_cpu') + backend_test.exclude(r'test_lppool_2d_default_cpu') + backend_test.exclude(r'test_lppool_2d_dilations_cpu') + backend_test.exclude(r'test_lppool_2d_pads_cpu') + backend_test.exclude(r'test_lppool_2d_same_lower_cpu') + backend_test.exclude(r'test_lppool_2d_same_upper_cpu') + backend_test.exclude(r'test_lppool_2d_strides_cpu') + backend_test.exclude(r'test_lppool_3d_default_cpu') + backend_test.exclude( + r'test_resize_downsample_scales_linear_half_pixel_symmetric_cpu') + backend_test.exclude( + r'test_resize_upsample_scales_linear_half_pixel_symmetric_cpu') + + # The following tests fail due to the CastLike operator being unsupported + backend_test.exclude(r'test_softplus_example_expanded_ver18_cpu') + backend_test.exclude(r'test_softplus_expanded_ver18_cpu') + backend_test.exclude(r'test_split_to_sequence_1_cpu') + backend_test.exclude(r'test_split_to_sequence_2_cpu') + backend_test.exclude(r'test_split_to_sequence_nokeepdims_cpu') + backend_test.exclude(r'test_wrap_pad_cpu') + + +def disabled_tests_float8(backend_test): + # e4m3fn (Prototensor data type 17 not supported) + backend_test.exclude(r'test_dequantizelinear_e4m3fn_cpu') + backend_test.exclude(r'test_quantizelinear_e4m3fn_cpu') + backend_test.exclude(r'test_cast_FLOAT16_to_FLOAT8E4M3FN_cpu') + backend_test.exclude(r'test_cast_FLOAT8E4M3FN_to_FLOAT16_cpu') + backend_test.exclude(r'test_cast_FLOAT8E4M3FN_to_FLOAT_cpu') + backend_test.exclude(r'test_cast_FLOAT_to_FLOAT8E4M3FN_cpu') + backend_test.exclude(r'test_cast_no_saturate_FLOAT16_to_FLOAT8E4M3FN_cpu') + backend_test.exclude(r'test_cast_no_saturate_FLOAT_to_FLOAT8E4M3FN_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E4M3FN_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E4M3FN_expanded_cpu') + # e4m3fnuz (Prototensor data type 18 not supported) + backend_test.exclude(r'test_cast_FLOAT16_to_FLOAT8E4M3FNUZ_cpu') + backend_test.exclude(r'test_cast_FLOAT8E4M3FNUZ_to_FLOAT16_cpu') + backend_test.exclude(r'test_cast_FLOAT8E4M3FNUZ_to_FLOAT_cpu') + backend_test.exclude(r'test_cast_FLOAT_to_FLOAT8E4M3FNUZ_cpu') + backend_test.exclude( + r'test_cast_no_saturate_FLOAT16_to_FLOAT8E4M3FNUZ_cpu') + backend_test.exclude(r'test_cast_no_saturate_FLOAT_to_FLOAT8E4M3FNUZ_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E4M3FNUZ_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E4M3FNUZ_to_FLOAT_expanded_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E4M3FN_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E4M3FN_to_FLOAT_expanded_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E4M3FNUZ_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E4M3FNUZ_expanded_cpu') + # e5m2 ( Prototensor data type 19 not supported ) + backend_test.exclude(r'test_dequantizelinear_e5m2_cpu') + backend_test.exclude(r'test_quantizelinear_e5m2_cpu') + backend_test.exclude(r'test_cast_FLOAT16_to_FLOAT8E5M2_cpu') + backend_test.exclude(r'test_cast_FLOAT8E5M2_to_FLOAT16_cpu') + backend_test.exclude(r'test_cast_FLOAT8E5M2_to_FLOAT_cpu') + backend_test.exclude(r'test_cast_FLOAT_to_FLOAT8E5M2_cpu') + backend_test.exclude(r'test_cast_no_saturate_FLOAT16_to_FLOAT8E5M2_cpu') + backend_test.exclude(r'test_cast_no_saturate_FLOAT_to_FLOAT8E5M2_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E5M2_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E5M2_expanded_cpu') + # e5m2fnuz (Prototensor data type 20 not supported) + backend_test.exclude(r'test_cast_FLOAT16_to_FLOAT8E5M2FNUZ_cpu') + backend_test.exclude(r'test_cast_FLOAT8E5M2FNUZ_to_FLOAT16_cpu') + backend_test.exclude(r'test_cast_FLOAT8E5M2FNUZ_to_FLOAT_cpu') + backend_test.exclude(r'test_cast_FLOAT_to_FLOAT8E5M2FNUZ_cpu') + backend_test.exclude( + r'test_cast_no_saturate_FLOAT16_to_FLOAT8E5M2FNUZ_cpu') + backend_test.exclude(r'test_cast_no_saturate_FLOAT_to_FLOAT8E5M2FNUZ_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E5M2FNUZ_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E5M2FNUZ_to_FLOAT_expanded_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E5M2_to_FLOAT_cpu') + backend_test.exclude(r'test_castlike_FLOAT8E5M2_to_FLOAT_expanded_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E5M2FNUZ_cpu') + backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT8E5M2FNUZ_expanded_cpu') + + +def disabled_tests_dynamic_shape(backend_test): + # constantofshape + backend_test.exclude(r'test_constantofshape_float_ones_cpu') + backend_test.exclude(r'test_constantofshape_int_shape_zero_cpu') + backend_test.exclude(r'test_constantofshape_int_zeros_cpu') + # cumsum + backend_test.exclude(r'test_cumsum_1d_cpu') + backend_test.exclude(r'test_cumsum_1d_exclusive_cpu') + backend_test.exclude(r'test_cumsum_1d_reverse_cpu') + backend_test.exclude(r'test_cumsum_1d_reverse_exclusive_cpu') + backend_test.exclude(r'test_cumsum_2d_axis_0_cpu') + backend_test.exclude(r'test_cumsum_2d_axis_1_cpu') + backend_test.exclude(r'test_cumsum_2d_negative_axis_cpu') + # expand + backend_test.exclude(r'test_expand_dim_changed_cpu') + backend_test.exclude(r'test_expand_dim_unchanged_cpu') + backend_test.exclude(r'test_expand_shape_model1_cpu') + backend_test.exclude(r'test_expand_shape_model2_cpu') + backend_test.exclude(r'test_expand_shape_model3_cpu') + backend_test.exclude(r'test_expand_shape_model4_cpu') + # onehot + backend_test.exclude(r'test_onehot_negative_indices_cpu') + backend_test.exclude(r'test_onehot_with_axis_cpu') + backend_test.exclude(r'test_onehot_with_negative_axis_cpu') + backend_test.exclude(r'test_onehot_without_axis_cpu') + # range + backend_test.exclude(r'test_range_float_type_positive_delta_cpu') + backend_test.exclude(r'test_range_int32_type_negative_delta_cpu') + # split + backend_test.exclude(r'test_split_variable_parts_1d_opset13_cpu') + backend_test.exclude(r'test_split_variable_parts_1d_opset18_cpu') + backend_test.exclude(r'test_split_variable_parts_2d_opset13_cpu') + backend_test.exclude(r'test_split_variable_parts_2d_opset18_cpu') + backend_test.exclude(r'test_split_variable_parts_default_axis_opset13_cpu') + backend_test.exclude(r'test_split_variable_parts_default_axis_opset18_cpu') + backend_test.exclude(r'test_split_zero_size_splits_opset13_cpu') + backend_test.exclude(r'test_split_zero_size_splits_opset18_cpu') + # squeeze + backend_test.exclude(r'test_squeeze_cpu') + backend_test.exclude(r'test_squeeze_negative_axes_cpu') + # unsqueeze + backend_test.exclude(r'test_unsqueeze_axis_0_cpu') + backend_test.exclude(r'test_unsqueeze_axis_1_cpu') + backend_test.exclude(r'test_unsqueeze_axis_2_cpu') + backend_test.exclude(r'test_unsqueeze_negative_axes_cpu') + backend_test.exclude(r'test_unsqueeze_three_axes_cpu') + backend_test.exclude(r'test_unsqueeze_two_axes_cpu') + backend_test.exclude(r'test_unsqueeze_unsorted_axes_cpu') + # tile + backend_test.exclude(r'test_tile_cpu') + backend_test.exclude(r'test_tile_precomputed_cpu') + # reshape + backend_test.exclude(r'test_reshape_allowzero_reordered_cpu') + backend_test.exclude(r'test_reshape_extended_dims_cpu') + backend_test.exclude(r'test_reshape_negative_dim_cpu') + backend_test.exclude(r'test_reshape_negative_extended_dims_cpu') + backend_test.exclude(r'test_reshape_one_dim_cpu') + backend_test.exclude(r'test_reshape_reduced_dims_cpu') + backend_test.exclude(r'test_reshape_reordered_all_dims_cpu') + backend_test.exclude(r'test_reshape_reordered_last_dims_cpu') + backend_test.exclude(r'test_reshape_zero_and_negative_dim_cpu') + backend_test.exclude(r'test_reshape_zero_dim_cpu') + # reduce backend_test.exclude( r'test_reduce_l1_default_axes_keepdims_example_expanded_cpu') - backend_test.exclude(r'test_reduce_l1_default_axes_keepdims_random_cpu') backend_test.exclude( r'test_reduce_l1_default_axes_keepdims_random_expanded_cpu') backend_test.exclude( @@ -174,10 +1053,8 @@ def disabled_tests_onnx_1_13_0(backend_test): r'test_reduce_l1_negative_axes_keep_dims_example_expanded_cpu') backend_test.exclude( r'test_reduce_l1_negative_axes_keep_dims_random_expanded_cpu') - backend_test.exclude(r'test_reduce_l2_default_axes_keepdims_example_cpu') backend_test.exclude( r'test_reduce_l2_default_axes_keepdims_example_expanded_cpu') - backend_test.exclude(r'test_reduce_l2_default_axes_keepdims_random_cpu') backend_test.exclude( r'test_reduce_l2_default_axes_keepdims_random_expanded_cpu') backend_test.exclude( @@ -195,12 +1072,8 @@ def disabled_tests_onnx_1_13_0(backend_test): backend_test.exclude(r'test_reduce_log_sum_default_expanded_cpu') backend_test.exclude(r'test_reduce_log_sum_desc_axes_cpu') backend_test.exclude(r'test_reduce_log_sum_desc_axes_expanded_cpu') - backend_test.exclude( - r'test_reduce_log_sum_exp_default_axes_keepdims_example_cpu') backend_test.exclude( r'test_reduce_log_sum_exp_default_axes_keepdims_example_expanded_cpu') - backend_test.exclude( - r'test_reduce_log_sum_exp_default_axes_keepdims_random_cpu') backend_test.exclude( r'test_reduce_log_sum_exp_default_axes_keepdims_random_expanded_cpu') backend_test.exclude( @@ -245,12 +1118,17 @@ def disabled_tests_onnx_1_13_0(backend_test): backend_test.exclude( r'test_reduce_prod_negative_axes_keepdims_example_cpu') backend_test.exclude(r'test_reduce_prod_negative_axes_keepdims_random_cpu') - backend_test.exclude( - r'test_reduce_sum_square_default_axes_keepdims_example_cpu') + backend_test.exclude(r'test_reduce_sum_default_axes_keepdims_example_cpu') + backend_test.exclude(r'test_reduce_sum_default_axes_keepdims_random_cpu') + backend_test.exclude(r'test_reduce_sum_do_not_keepdims_example_cpu') + backend_test.exclude(r'test_reduce_sum_do_not_keepdims_random_cpu') + backend_test.exclude(r'test_reduce_sum_empty_axes_input_noop_example_cpu') + backend_test.exclude(r'test_reduce_sum_keepdims_example_cpu') + backend_test.exclude(r'test_reduce_sum_keepdims_random_cpu') + backend_test.exclude(r'test_reduce_sum_negative_axes_keepdims_example_cpu') + backend_test.exclude(r'test_reduce_sum_negative_axes_keepdims_random_cpu') backend_test.exclude( r'test_reduce_sum_square_default_axes_keepdims_example_expanded_cpu') - backend_test.exclude( - r'test_reduce_sum_square_default_axes_keepdims_random_cpu') backend_test.exclude( r'test_reduce_sum_square_default_axes_keepdims_random_expanded_cpu') backend_test.exclude( @@ -265,47 +1143,6 @@ def disabled_tests_onnx_1_13_0(backend_test): r'test_reduce_sum_square_negative_axes_keepdims_example_expanded_cpu') backend_test.exclude( r'test_reduce_sum_square_negative_axes_keepdims_random_expanded_cpu') - backend_test.exclude(r'test_scatter_elements_with_reduction_max_cpu') - backend_test.exclude(r'test_scatter_elements_with_reduction_min_cpu') - - # The following tests fail due to the CastLike operator being unsupported - backend_test.exclude(r'test_elu_default_expanded_ver18_cpu') - backend_test.exclude(r'test_elu_example_expanded_ver18_cpu') - backend_test.exclude(r'test_elu_expanded_ver18_cpu') - backend_test.exclude(r'test_hardsigmoid_default_expanded_ver18_cpu') - backend_test.exclude(r'test_hardsigmoid_example_expanded_ver18_cpu') - backend_test.exclude(r'test_hardsigmoid_expanded_ver18_cpu') - backend_test.exclude(r'test_leakyrelu_default_expanded_cpu') - backend_test.exclude(r'test_leakyrelu_example_expanded_cpu') - backend_test.exclude(r'test_leakyrelu_expanded_cpu') - backend_test.exclude(r'test_selu_default_expanded_ver18_cpu') - backend_test.exclude(r'test_selu_example_expanded_ver18_cpu') - backend_test.exclude(r'test_selu_expanded_ver18_cpu') - backend_test.exclude(r'test_thresholdedrelu_default_expanded_ver18_cpu') - backend_test.exclude(r'test_thresholdedrelu_example_expanded_ver18_cpu') - backend_test.exclude(r'test_thresholdedrelu_expanded_ver18_cpu') - backend_test.exclude(r'test_relu_expanded_ver18_cpu') - backend_test.exclude(r'test_softsign_example_expanded_ver18_cpu') - backend_test.exclude(r'test_softsign_expanded_ver18_cpu') - - -def disabled_tests_onnx_1_14_0(backend_test): - # fails - backend_test.exclude(r'test_averagepool_2d_dilations_cpu') - backend_test.exclude(r'test_roialign_mode_max_cpu') - - # errors - backend_test.exclude(r'test_constant_pad_negative_axes_cpu') - backend_test.exclude(r'test_dequantizelinear_e4m3fn_cpu') - backend_test.exclude(r'test_dequantizelinear_e5m2_cpu') - backend_test.exclude(r'test_equal_string_broadcast_cpu') - backend_test.exclude(r'test_equal_string_cpu') - backend_test.exclude(r'test_quantizelinear_e4m3fn_cpu') - backend_test.exclude(r'test_quantizelinear_e5m2_cpu') - - # The following tests fail due to the CastLike operator being unsupported - backend_test.exclude(r'test_softplus_example_expanded_ver18_cpu') - backend_test.exclude(r'test_softplus_expanded_ver18_cpu') def create_backend_test(testname=None, target_device=None): @@ -316,8 +1153,7 @@ def create_backend_test(testname=None, target_device=None): if testname: backend_test.include(testname + '.*') else: - # Include all of the nodes that we support. - # Onnx native node tests + # Onnx Operator tests backend_test.include(r'.*test_abs.*') backend_test.include(r'.*test_acos.*') backend_test.include(r'.*test_acosh.*') @@ -331,73 +1167,209 @@ def create_backend_test(testname=None, target_device=None): backend_test.include(r'.*test_atanh.*') backend_test.include(r'.*test_averagepool.*') backend_test.include(r'.*test_AvgPool.*') - backend_test.include(r'.*test_BatchNorm.*eval.*') + backend_test.include(r'.*test_[bB]atch[nN]orm(?!.*training).*') + backend_test.include(r'.*test_bitshift.*') + backend_test.include(r'.*test_bitwise.*') backend_test.include(r'.*test_ceil.*') - backend_test.include(r'.*test_celu.*') - backend_test.include(r'.*test_clip.*') + backend_test.include(r'.*test_cast_.*') + backend_test.include(r'.*test_col2im.*') + backend_test.include(r'.*test_compress.*') backend_test.include(r'.*test_concat.*') - backend_test.include(r'.*test_constant.*') - backend_test.include(r'.*test_Conv[1-3]d*') + backend_test.include(r'.*test_constant_.*') + backend_test.include(r'.*test_Constant.*') + backend_test.include(r'.*test_constantofshape.*') + backend_test.include(r'.*test_(basic_)?conv_.*') + backend_test.include(r'.*test_Conv[1-3]d.*') + backend_test.include(r'.*test_convinteger.*') + backend_test.include(r'.*test_convtranspose.*') + backend_test.include(r'.*test_ConvTranspose[1-3]d.*') backend_test.include(r'.*test_cos.*') backend_test.include(r'.*test_cosh.*') + backend_test.include(r'.*test_cumsum.*') + backend_test.include(r'.*test_(basic_)?deform_conv.*') backend_test.include(r'.*test_depthtospace.*') - backend_test.include(r'.*test_dequantizelinear') + backend_test.include(r'.*test_dequantizelinear.*') + backend_test.include(r'.*test_det.*') + backend_test.include(r'.*test_dft.*') backend_test.include(r'.*test_div.*') backend_test.include(r'.*test_dropout.*') - backend_test.include(r'.*test_ELU*') - backend_test.include(r'.*test_elu.*') + backend_test.include(r'.*test_einsum.*') backend_test.include(r'.*test_equal.*') - backend_test.include(r'.*test_Embedding*') - backend_test.include(r'.*test_exp.*') + backend_test.include(r'.*test_Embedding.*') + backend_test.include(r'.*test_erf.*') + backend_test.include(r'.*test_exp_.*') + backend_test.include(r'.*test_expand.*') backend_test.include(r'.*test_eyelike.*') backend_test.include(r'.*test_flatten.*') backend_test.include(r'.*test_floor.*') - backend_test.include(r'.*test_fmod.*') + backend_test.include(r'.*test_gru.*') backend_test.include(r'.*test_gather.*') backend_test.include(r'.*test_gemm.*') backend_test.include(r'.*test_globalaveragepool.*') + backend_test.include(r'.*test_globallppool.*') backend_test.include(r'.*test_globalmaxpool.*') backend_test.include(r'.*test_greater.*') - backend_test.include(r'.*test_hardsigmoid.*') - backend_test.include(r'.*test_hardswish.*') + backend_test.include(r'.*test_gridsample.*') + backend_test.include(r'.*test_hardmax.*') backend_test.include(r'.*test_identity.*') backend_test.include(r'.*test_if.*') + backend_test.include(r'.*test_instancenorm.*') + backend_test.include(r'.*test_isinf.*') backend_test.include(r'.*test_isnan.*') - backend_test.include(r'.*test_LeakyReLU*') - backend_test.include(r'.*test_leakyrelu.*') - backend_test.include(r'.*test_less.*') - backend_test.include(r'.*test_Linear.*') - backend_test.include(r'.*test_log.*') - backend_test.include(r'.*test_logsoftmax.*') - backend_test.include(r'.*test_LogSoftmax.*') - backend_test.include(r'.*test_log_softmax.*') backend_test.include(r'.*test_lrn.*') + backend_test.include(r'.*test_lstm.*') + backend_test.include(r'.*test_log.*') + backend_test.include(r'.*test_loop.*') + backend_test.include(r'.*test_lpnorm.*') + backend_test.include(r'.*test_lppool.*') backend_test.include(r'.*test_matmul.*') - backend_test.include(r'.*test_max.*') - backend_test.include(r'.*test_MaxPool[1-9]d.*') + backend_test.include(r'.*test_max_.*') + backend_test.include(r'.*test_maxpool.*') + backend_test.include(r'.*test_MaxPool[1-3]d.*') + backend_test.include(r'.*test_maxroipool.*') + backend_test.include(r'.*test_maxunpool.*') backend_test.include(r'.*test_mean.*') + backend_test.include(r'.*test_melweightmatrix.*') backend_test.include(r'.*test_min.*') - backend_test.include(r' .*test_mod.*') + backend_test.include(r'.*test_mod.*') backend_test.include(r'.*test_mul.*') - backend_test.include(r'.*test_multinomial.*') - backend_test.include(r'.*test_Multinomial.*') + backend_test.include(r'.*test_[mM]ultinomial.*') backend_test.include(r'.*test_neg.*') + backend_test.include(r'.*test_nonmaxsuppression.*') + backend_test.include(r'.*test_nonzero.*') backend_test.include(r'.*test_not.*') + backend_test.include(r'.*test_onehot.*') + backend_test.include(r'.*optional_get_element.*') + backend_test.include(r'.*optional_has_element.*') + backend_test.include(r'.*test_or.*') + backend_test.include(r'.*test_(constant_|edge_|reflect_|wrap_)?pad.*') + backend_test.include( + r'.*test_(Constant|Reflection|Replication|Zero)+Pad2d.*') + backend_test.include(r'.*test_pow.*') + backend_test.include(r'.*test_qlinearconv.*') + backend_test.include(r'.*test_qlinearmatmul.*') + backend_test.include(r'.*test_quantizelinear.*') + backend_test.include(r'.*test_(simple_)?rnn.*') + backend_test.include(r'.*test_randomnormal.*') + backend_test.include(r'.*test_randomuniform.*') + backend_test.include(r'.*test_reciprocal.*') + backend_test.include(r'.*test_reduce_max.*') + backend_test.include(r'.*test_reduce_mean.*') + backend_test.include(r'.*test_reduce_min.*') + backend_test.include(r'.*test_reduce_prod.*') + backend_test.include(r'.*test_reduce_sum.*') + backend_test.include(r'.*test_reshape.*') + backend_test.include(r'.*test_resize.*') + backend_test.include(r'.*test_reversesequence.*') + backend_test.include(r'.*test_roialign.*') + backend_test.include(r'.*test_round.*') + backend_test.include(r'.*test_stft.*') + backend_test.include(r'.*test_scan.*') + backend_test.include(r'.*test_scatter.*') + backend_test.include(r'.*test_sequence_at.*') + backend_test.include(r'.*test_sequence_construct.*') + backend_test.include(r'.*test_sequence_empty.*') + backend_test.include(r'.*test_sequence_erase.*') + backend_test.include(r'.*test_sequence_insert.*') + backend_test.include(r'.*test_sequence_length.*') + backend_test.include(r'.*test_shape.*') + backend_test.include(r'.*test_[sS]igmoid.*') + backend_test.include(r'.*test_sign.*') + backend_test.include(r'.*test_sin_.*') + backend_test.include(r'.*test_sinh.*') + backend_test.include(r'.*test_size.*') + backend_test.include(r'.*test_slice.*') + backend_test.include(r'.*test_spacetodepth.*') + backend_test.include(r'.*test_split.*') + backend_test.include(r'.*test_split_to_sequence.*') + backend_test.include(r'.*test_sqrt.*') + backend_test.include(r'.*test_squeeze.*') + backend_test.include(r'.*test_squeeze.*') + backend_test.include(r'.*test_strnorm.*') + backend_test.include(r'.*test_sub.*') + backend_test.include(r'.*test_sum.*') + backend_test.include(r'.*test_tan_.*') + backend_test.include(r'.*test_[tT]anh.*') + backend_test.include(r'.*test_tfidfvectorizer.*') + backend_test.include(r'.*test_tile.*') + backend_test.include(r'.*test_top_k.*') + backend_test.include(r'.*test_transpose.*') + backend_test.include(r'.*test_tril.*') + backend_test.include(r'.*test_triu.*') + backend_test.include(r'.*test_unique.*') + backend_test.include(r'.*test_unsqueeze.*') + backend_test.include(r'.*test_upsample.*') + backend_test.include(r'.*test_where.*') + backend_test.include(r'.*test_xor.*') + + # Onnx Function tests + backend_test.include(r'.*test_bernoulli.*') + backend_test.include(r'.*test_blackmanwindow.*') + backend_test.include(r'.*test_castlike.*') + backend_test.include(r'.*test_celu.*') + backend_test.include(r'.*test_center_crop_pad.*') + backend_test.include(r'.*test_clip.*') + backend_test.include(r'.*test_dynamicquantizelinear.*') + backend_test.include(r'.*test_elu.*') + backend_test.include(r'.*test_ELU.*') + backend_test.include(r'.*test_GLU.*') + backend_test.include(r'.*test_greater_equal.*') + backend_test.include(r'.*test_group_normalization.*') + backend_test.include(r'.*test_hammingwindow.*') + backend_test.include(r'.*test_hannwindow.*') + backend_test.include(r'.*test_hardsigmoid.*') + backend_test.include(r'.*test_hardswish.*') + backend_test.include(r'.*test_layer_normalization.*') + backend_test.include(r'.*test_LeakyReLU.*') + backend_test.include(r'.*test_leakyrelu.*') + backend_test.include(r'.*test_less.*') + backend_test.include(r'.*test_Linear.*') + backend_test.include(r'.*test_logsoftmax.*') + backend_test.include(r'.*test_log_softmax.*') + backend_test.include(r'.*test_LogSoftmax.*') + backend_test.include(r'.*test_mvn.*') + backend_test.include(r'.*test_mish.*') + backend_test.include(r'.*test_nllloss.*') + backend_test.include(r'.*test_PixelShuffle.*') + backend_test.include(r'.*test_PoissonNLLLLoss_no_reduce.*') + backend_test.include(r'.*test_prelu.*') + backend_test.include(r'.*test_PReLU.*') + backend_test.include(r'.*test_range.*') + backend_test.include(r'.*test_reduce_l1.*') + backend_test.include(r'.*test_reduce_l2.*') + backend_test.include(r'.*test_reduce_log.*') + backend_test.include(r'.*test_ReLU.*') + backend_test.include(r'.*test_relu.*') + backend_test.include(r'.*test_selu.*') + backend_test.include(r'.*test_SELU.*') + backend_test.include(r'.*test_sequence_map.*') + backend_test.include(r'.*test_shrink.*') + backend_test.include(r'.*test_[sS]oftmax.*') + backend_test.include(r'.*test_[sS]oftplus.*') + backend_test.include(r'.*test_[sS]oftsign.*') + backend_test.include(r'.*test_sce.*') + backend_test.include(r'.*test_thresholdedrelu.*') + + # OnnxBackendPyTorchOperatorModelTest + backend_test.include(r'.*test_operator_add_broadcast.*') + backend_test.include(r'.*test_operator_addconstant.*') backend_test.include(r'.*test_operator_addmm.*') + backend_test.include(r'.*test_operator_add_size1.*') backend_test.include(r'.*test_operator_basic.*') backend_test.include(r'.*test_operator_chunk.*') backend_test.include(r'.*test_operator_clip.*') backend_test.include(r'.*test_operator_concat2.*') backend_test.include(r'.*test_operator_conv_.*') + backend_test.include(r'.*test_operator_convtranspose.*') backend_test.include(r'.*test_operator_exp.*') backend_test.include(r'.*test_operator_flatten.*') backend_test.include(r'.*test_operator_index.*') backend_test.include(r'.*test_operator_max_.*') backend_test.include(r'.*test_operator_maxpool.*') backend_test.include(r'.*test_operator_min.*') - backend_test.include(r'.*test_operator_mod.*') backend_test.include(r'.*test_operator_mm.*') backend_test.include(r'.*test_operator_non_float_params.*') + backend_test.include(r'.*test_operator_pad.*') backend_test.include(r'.*test_operator_params.*') backend_test.include(r'.*test_operator_permute2.*') backend_test.include(r'.*test_operator_pow.*') @@ -405,55 +1377,19 @@ def create_backend_test(testname=None, target_device=None): backend_test.include(r'.*test_operator_reduced_mean_keepdim.*') backend_test.include(r'.*test_operator_reduced_sum_.*') backend_test.include(r'.*test_operator_reduced_sum_keepdim.*') + backend_test.include(r'.*test_operator_repeat.*') backend_test.include(r'.*test_operator_selu.*') backend_test.include(r'.*test_operator_sqrt.*') backend_test.include(r'.*test_operator_symbolic_override.*') backend_test.include(r'.*test_operator_symbolic_override_nested.*') backend_test.include(r'.*test_operator_view.*') - backend_test.include(r'.*test_or.*') - backend_test.include(r'.*test_pow.*') - backend_test.include(r'.*test_PoissonNLLLLoss_no_reduce*') - backend_test.include(r'.*test_quantizelinear') - backend_test.include(r'.*test_reciprocal.*') - backend_test.include(r'.*test_reduce.*') - backend_test.include(r'.*test_ReLU*') - backend_test.include(r'.*test_relu.*') - #backend_test.include(r'.*test_reversesequence.*') - backend_test.include(r'.*test_RoiAlign*') - backend_test.include(r'.*test_roialign.*') - backend_test.include(r'.*test_scatter.*') - backend_test.include(r'.*test_Scatter.*') - backend_test.include(r'.*test_selu.*') - backend_test.include(r'.*test_shape.*') - backend_test.include(r'.*test_Sigmoid*') - backend_test.include(r'.*test_sigmoid.*') - backend_test.include(r'.*test_sin.*') - backend_test.include(r'.*test_sinh.*') - backend_test.include(r'.*test_size.*') - backend_test.include(r'.*test_Softmax*') - backend_test.include(r'.*test_softmax.*') - backend_test.include(r'.*test_Softmin*') - backend_test.include(r'.*test_Softplus*') - backend_test.include(r'.*test_softplus.*') - backend_test.include(r'.*test_softsign.*') - backend_test.include(r'.*test_sqrt.*') - backend_test.include(r'.*test_squeeze_cuda') - backend_test.include(r'.*test_sub.*') - backend_test.include(r'.*test_sum.*') - backend_test.include(r'.*test_tan.*') - backend_test.include(r'.*test_Tanh*') - backend_test.include(r'.*test_tanh.*') - backend_test.include(r'.*test_thresholdedrelu.*') - backend_test.include(r'.*test_topk.*') - backend_test.include(r'.*test_Topk.*') - backend_test.include(r'.*test_transpose.*') - backend_test.include(r'.*test_unsqueeze.*') - backend_test.include(r'.*test_where*') - backend_test.include(r'.*test_where.*') - backend_test.include(r'.*test_xor.*') - backend_test.include(r'.*test_ZeroPad2d*') - # # Onnx native model tests + # OnnxBackendSimpleModelTest + backend_test.include(r'.*test_gradient_of.*') + backend_test.include(r'.*test_sequence_model.*') + backend_test.include(r'.*test_single_relu_model.*') + + # OnnxBackendRealModelTest backend_test.include(r'.*test_bvlc_alexnet.*') backend_test.include(r'.*test_densenet121.*') backend_test.include(r'.*test_inception_v1.*') @@ -464,69 +1400,42 @@ def create_backend_test(testname=None, target_device=None): backend_test.include(r'.*test_vgg19.*') backend_test.include(r'.*test_zfnet512.*') - # exclude unenabled ops get pulled in with wildcards - # test_constant_pad gets pulled in with the test_constant* wildcard. Explicitly disable padding tests for now. - # Operator MATMULINTEGER is not supported by TRT - backend_test.exclude(r'.*test_matmulinteger.*') - backend_test.exclude(r'.*test_maxunpool.*') - # Absolute diff failed because - # numpy compares the difference between actual and desired to atol + rtol * abs(desired) - - # failed test cases - backend_test.exclude( - r'test_argmax_keepdims_example_select_last_index_cpu') - backend_test.exclude( - r'test_argmax_negative_axis_keepdims_example_select_last_index_cpu' - ) - backend_test.exclude( - r'test_argmax_no_keepdims_example_select_last_index_cpu') - backend_test.exclude( - r'test_argmin_keepdims_example_select_last_index_cpu') - backend_test.exclude( - r'test_argmin_negative_axis_keepdims_example_select_last_index_cpu' - ) - backend_test.exclude( - r'test_argmin_no_keepdims_example_select_last_index_cpu') - backend_test.exclude(r'test_lrn_cpu') - backend_test.exclude(r'test_lrn_default_cpu') - backend_test.exclude(r'test_maxpool_2d_dilations_cpu') - backend_test.exclude(r'test_MaxPool2d_stride_padding_dilation_cpu') - backend_test.exclude(r'test_MaxPool1d_stride_padding_dilation_cpu') - backend_test.exclude( - r'test_maxpool_with_argmax_2d_precomputed_pads_cpu') - backend_test.exclude( - r'test_maxpool_with_argmax_2d_precomputed_strides_cpu') - - # error cases - backend_test.exclude(r'test_constant_pad_cpu') - backend_test.exclude(r'test_constantofshape_float_ones_cpu') - backend_test.exclude(r'test_constantofshape_int_shape_zero_cpu') - backend_test.exclude(r'test_constantofshape_int_zeros_cpu') - backend_test.exclude(r'test_expand_dim_changed_cpu') - backend_test.exclude(r'test_expand_dim_unchanged_cpu') - backend_test.exclude(r'test_expand_shape_model1_cpu') - backend_test.exclude(r'test_expand_shape_model2_cpu') - backend_test.exclude(r'test_expand_shape_model3_cpu') - backend_test.exclude(r'test_expand_shape_model4_cpu') - backend_test.exclude(r'test_identity_sequence_cpu') - backend_test.exclude(r'test_maxpool_2d_uint8_cpu') - backend_test.exclude(r'test_negative_log_likelihood_loss_*') - - # all reduce ops have dynamic axes inputs - backend_test.exclude(r'test_softmax_cross_entropy_*') - backend_test.exclude(r'test_Embedding_cpu') - - # real model tests + # Skipped tests + # backend_test.include(r'.*test_adagrad.*') + # backend_test.include(r'.*test_adam.*') + # backend_test.include(r'.*test_ai_onnx_ml.*') + # backend_test.include(r'.*test_batchnorm_epsilon_training.*') + # backend_test.include(r'.*test_batchnorm_example_training.*') + # backend_test.include(r'.*test_momentum.*') + # backend_test.include(r'.*test_nesterov_momentum.*') + # backend_test.include(r'.*test_training_dropout.*') + # backend_test.include(r'.*test_Softmin.*') + + # Exclude failing tests + + # from OnnxBackendRealModelTest backend_test.exclude(r'test_inception_v1_cpu') backend_test.exclude(r'test_resnet50_cpu') backend_test.exclude(r'test_squeezenet_cpu') + # PRelu OnnxBackendPyTorchConvertedModelTest has wrong dim for broadcasting + backend_test.exclude(r'[a-z,_]*PReLU_[0-9]d_multiparam[a-z,_]*') + + # Remove when float8 is supported + disabled_tests_float8(backend_test) + + # Remove when dynamic shapes are supported + disabled_tests_dynamic_shape(backend_test) + # additional cases disabled for a specific onnx version - if version.parse(onnx.__version__) <= version.parse("1.7.0"): + if version.parse(onnx.__version__) >= version.parse("1.7.0"): disabled_tests_onnx_1_7_0(backend_test) if version.parse(onnx.__version__) >= version.parse("1.8.0"): - disabled_tests_onnx_1_8_1(backend_test) + disabled_tests_onnx_1_8_0(backend_test) + + if version.parse(onnx.__version__) >= version.parse("1.9.0"): + disabled_tests_onnx_1_9_0(backend_test) if version.parse(onnx.__version__) >= version.parse("1.10.0"): disabled_tests_onnx_1_10_0(backend_test) From bd425d0d1de6f6c779226d7a985ab9c0a6661e83 Mon Sep 17 00:00:00 2001 From: Attila Dusnoki <126579622+attila-dusnoki-htec@users.noreply.github.com> Date: Tue, 3 Oct 2023 20:54:29 +0200 Subject: [PATCH 08/19] Remove onnx_backend_test.py from top level cmake (#2256) --- CMakeLists.txt | 1 - 1 file changed, 1 deletion(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index 499e307bc37..b99f97d4027 100755 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -283,7 +283,6 @@ file(MAKE_DIRECTORY ${DEST_DIR}/lib/onnx_migraphx) foreach(py_file ${backend_files}) configure_file(${py_file} ${DEST_DIR}/lib/onnx_migraphx/. COPYONLY) endforeach(py_file) -configure_file(${CMAKE_SOURCE_DIR}/test/py/onnx_backend_test.py ${DEST_DIR}/onnx_backend_test.py COPYONLY) rocm_create_package( NAME MIGraphX From b72cae6d1e0590f86d4215fdc87df8ed955c21b0 Mon Sep 17 00:00:00 2001 From: Chris Austen Date: Tue, 3 Oct 2023 18:34:18 -0400 Subject: [PATCH 09/19] update version after branch (#2273) --- CMakeLists.txt | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/CMakeLists.txt b/CMakeLists.txt index b99f97d4027..58e1dc9a6d3 100755 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -71,7 +71,7 @@ include(ROCMSetupVersion) option(BUILD_DEV "Build for development purpose only" OFF) -rocm_setup_version(VERSION 2.7.0) +rocm_setup_version(VERSION 2.8.0) set(MIGRAPHX_SO_VERSION ${PROJECT_VERSION_MAJOR}.${PROJECT_VERSION_MINOR}.${PROJECT_VERSION_PATCH}) option( BUILD_SHARED_LIBS "Build as a shared library" ON ) From 9660c64c178cf42a479586b1c3ec6c502ba20cab Mon Sep 17 00:00:00 2001 From: Umang Yadav <29876643+umangyadav@users.noreply.github.com> Date: Tue, 3 Oct 2023 19:29:03 -0400 Subject: [PATCH 10/19] Avoid using Template argument deduction for Vector (#2276) --- .github/workflows/ci.yaml | 1 + src/onnx/parse_resize.cpp | 3 ++- 2 files changed, 3 insertions(+), 1 deletion(-) diff --git a/.github/workflows/ci.yaml b/.github/workflows/ci.yaml index aad99c2cd1f..0cd5d708749 100644 --- a/.github/workflows/ci.yaml +++ b/.github/workflows/ci.yaml @@ -281,6 +281,7 @@ jobs: -DBUILD_DEV=On \ -DCMAKE_CXX_COMPILER_LAUNCHER=/usr/local/bin/ccache \ -DCMAKE_C_COMPILER_LAUNCHER=/usr/local/bin/ccache \ + -DCMAKE_CXX_FLAGS="-Werror" \ -DGPU_TARGETS=gfx908 \ .. make -j$(nproc) tests driver diff --git a/src/onnx/parse_resize.cpp b/src/onnx/parse_resize.cpp index 13fd4760e22..93ad95bdb27 100644 --- a/src/onnx/parse_resize.cpp +++ b/src/onnx/parse_resize.cpp @@ -320,7 +320,8 @@ struct parse_resize : op_parser // get the number of dimensions std::size_t n_dim = out_lens.size(); - auto vvv_ind = std::vector(n_dim, std::vector(2, std::vector(out_elements))); + std::vector> vv_ind(2, std::vector(out_elements)); + std::vector>> vvv_ind(n_dim, vv_ind); std::vector> delta(n_dim, std::vector(out_elements)); shape_for_each(out_s, [&](const auto& out_idx_v, size_t out_idx) { From 42a7e55da831b9e40cb1d1c4395e54807156e710 Mon Sep 17 00:00:00 2001 From: Zakor Gyula <126694206+gyulaz-htec@users.noreply.github.com> Date: Wed, 4 Oct 2023 14:42:09 +0200 Subject: [PATCH 11/19] Add support for CastLike operator (#2261) --- src/onnx/parse_castlike.cpp | 53 ++++++++++++++++++++++++++++++ test/onnx/castlike_error_test.onnx | 14 ++++++++ test/onnx/castlike_test.onnx | 20 +++++++++++ test/onnx/gen_onnx.py | 23 +++++++++++++ test/onnx/onnx_test.cpp | 20 +++++++++++ test/py/onnx_backend_test.py | 26 --------------- 6 files changed, 130 insertions(+), 26 deletions(-) create mode 100644 src/onnx/parse_castlike.cpp create mode 100644 test/onnx/castlike_error_test.onnx create mode 100644 test/onnx/castlike_test.onnx diff --git a/src/onnx/parse_castlike.cpp b/src/onnx/parse_castlike.cpp new file mode 100644 index 00000000000..61d1fb27b05 --- /dev/null +++ b/src/onnx/parse_castlike.cpp @@ -0,0 +1,53 @@ +/* + * The MIT License (MIT) + * + * Copyright (c) 2015-2023 Advanced Micro Devices, Inc. All rights reserved. + * + * Permission is hereby granted, free of charge, to any person obtaining a copy + * of this software and associated documentation files (the "Software"), to deal + * in the Software without restriction, including without limitation the rights + * to use, copy, modify, merge, publish, distribute, sublicense, and/or sell + * copies of the Software, and to permit persons to whom the Software is + * furnished to do so, subject to the following conditions: + * + * The above copyright notice and this permission notice shall be included in + * all copies or substantial portions of the Software. + * + * THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR + * IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY, + * FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE + * AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER + * LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM, + * OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN + * THE SOFTWARE. + */ +#include +#include +#include +#include + +namespace migraphx { +inline namespace MIGRAPHX_INLINE_NS { +namespace onnx { + +struct parse_castlike : op_parser +{ + std::vector operators() const { return {{"CastLike"}}; } + + instruction_ref parse(const op_desc& /*opd*/, + const onnx_parser& /*parser*/, + const onnx_parser::node_info& info, + const std::vector& args) const + { + if(not(args.size() == 2)) + { + MIGRAPHX_THROW("PARSE_CASTLIKE: CastLike must have exactly 2 inputs!"); + } + shape::type_t target_type = args[1]->get_shape().type(); + return info.add_instruction(make_op("convert", {{"target_type", target_type}}), args[0]); + } +}; + +} // namespace onnx +} // namespace MIGRAPHX_INLINE_NS +} // namespace migraphx diff --git a/test/onnx/castlike_error_test.onnx b/test/onnx/castlike_error_test.onnx new file mode 100644 index 00000000000..8d50faef0f1 --- /dev/null +++ b/test/onnx/castlike_error_test.onnx @@ -0,0 +1,14 @@ + castlike_error_test:M + +0out"CastLikecastlike_error_testZ +0 + + + + +b +out + + + +B \ No newline at end of file diff --git a/test/onnx/castlike_test.onnx b/test/onnx/castlike_test.onnx new file mode 100644 index 00000000000..f7066991c2b --- /dev/null +++ b/test/onnx/castlike_test.onnx @@ -0,0 +1,20 @@ +  castlike_test:[ + +0 +1out"CastLike castlike_testZ +0 + + + + +Z +1 + + + +b +out + + + +B \ No newline at end of file diff --git a/test/onnx/gen_onnx.py b/test/onnx/gen_onnx.py index d29ca03e9b9..b1dda835829 100644 --- a/test/onnx/gen_onnx.py +++ b/test/onnx/gen_onnx.py @@ -582,6 +582,29 @@ def cast_test(): return ([node], [x], [y]) +@onnx_test() +def castlike_test(): + input = helper.make_tensor_value_info('0', TensorProto.FLOAT16, [10]) + target_type = helper.make_tensor_value_info('1', TensorProto.FLOAT, [10]) + output = helper.make_tensor_value_info('out', TensorProto.FLOAT, [10]) + + node = onnx.helper.make_node('CastLike', + inputs=['0', '1'], + outputs=['out']) + + return ([node], [input, target_type], [output]) + + +@onnx_test() +def castlike_error_test(): + input = helper.make_tensor_value_info('0', TensorProto.FLOAT16, [10]) + output = helper.make_tensor_value_info('out', TensorProto.FLOAT, [10]) + + node = onnx.helper.make_node('CastLike', inputs=['0'], outputs=['out']) + + return ([node], [input], [output]) + + @onnx_test() def ceil_test(): x = helper.make_tensor_value_info('x', TensorProto.FLOAT, [10]) diff --git a/test/onnx/onnx_test.cpp b/test/onnx/onnx_test.cpp index 8c545bf0b3b..493b4aee308 100644 --- a/test/onnx/onnx_test.cpp +++ b/test/onnx/onnx_test.cpp @@ -687,6 +687,26 @@ TEST_CASE(cast_test) EXPECT(p == prog); } +TEST_CASE(castlike_test) +{ + migraphx::program p; + auto* mm = p.get_main_module(); + auto l = mm->add_parameter("0", migraphx::shape{migraphx::shape::half_type, {10}}); + mm->add_parameter("1", migraphx::shape{migraphx::shape::float_type, {10}}); + mm->add_instruction( + migraphx::make_op("convert", + {{"target_type", migraphx::to_value(migraphx::shape::float_type)}}), + l); + + auto prog = optimize_onnx("castlike_test.onnx"); + EXPECT(p == prog); +} + +TEST_CASE(castlike_error_test) +{ + EXPECT(test::throws([&] { migraphx::parse_onnx("castlike_error_test.onnx"); })); +} + TEST_CASE(ceil_test) { migraphx::program p; diff --git a/test/py/onnx_backend_test.py b/test/py/onnx_backend_test.py index 30f8c53ab6e..3bafd6f2cae 100644 --- a/test/py/onnx_backend_test.py +++ b/test/py/onnx_backend_test.py @@ -635,14 +635,8 @@ def disabled_tests_onnx_1_10_0(backend_test): backend_test.exclude(r'test_bernoulli_seed_cpu') backend_test.exclude(r'test_castlike_BFLOAT16_to_FLOAT_cpu') backend_test.exclude(r'test_castlike_BFLOAT16_to_FLOAT_expanded_cpu') - backend_test.exclude(r'test_castlike_DOUBLE_to_FLOAT16_cpu') - backend_test.exclude(r'test_castlike_DOUBLE_to_FLOAT_cpu') - backend_test.exclude(r'test_castlike_FLOAT16_to_DOUBLE_cpu') - backend_test.exclude(r'test_castlike_FLOAT16_to_FLOAT_cpu') backend_test.exclude(r'test_castlike_FLOAT_to_BFLOAT16_cpu') backend_test.exclude(r'test_castlike_FLOAT_to_BFLOAT16_expanded_cpu') - backend_test.exclude(r'test_castlike_FLOAT_to_DOUBLE_cpu') - backend_test.exclude(r'test_castlike_FLOAT_to_FLOAT16_cpu') backend_test.exclude(r'test_castlike_FLOAT_to_STRING_cpu') backend_test.exclude(r'test_castlike_FLOAT_to_STRING_expanded_cpu') backend_test.exclude(r'test_castlike_STRING_to_FLOAT_cpu') @@ -860,28 +854,8 @@ def disabled_tests_onnx_1_13_0(backend_test): backend_test.exclude(r'test_scatter_elements_with_reduction_min_cpu') # The following tests fail due to the CastLike operator being unsupported - backend_test.exclude(r'test_elu_default_expanded_ver18_cpu') - backend_test.exclude(r'test_elu_example_expanded_ver18_cpu') - backend_test.exclude(r'test_elu_expanded_ver18_cpu') - backend_test.exclude(r'test_hardsigmoid_default_expanded_ver18_cpu') - backend_test.exclude(r'test_hardsigmoid_example_expanded_ver18_cpu') - backend_test.exclude(r'test_hardsigmoid_expanded_ver18_cpu') - backend_test.exclude(r'test_leakyrelu_default_expanded_cpu') - backend_test.exclude(r'test_leakyrelu_example_expanded_cpu') - backend_test.exclude(r'test_leakyrelu_expanded_cpu') - backend_test.exclude(r'test_selu_default_expanded_ver18_cpu') - backend_test.exclude(r'test_selu_example_expanded_ver18_cpu') - backend_test.exclude(r'test_selu_expanded_ver18_cpu') - backend_test.exclude(r'test_shrink_hard_expanded_ver18_cpu') - backend_test.exclude(r'test_shrink_soft_expanded_ver18_cpu') backend_test.exclude(r'test_split_1d_uneven_split_opset18_cpu') backend_test.exclude(r'test_split_2d_uneven_split_opset18_cpu') - backend_test.exclude(r'test_thresholdedrelu_default_expanded_ver18_cpu') - backend_test.exclude(r'test_thresholdedrelu_example_expanded_ver18_cpu') - backend_test.exclude(r'test_thresholdedrelu_expanded_ver18_cpu') - backend_test.exclude(r'test_relu_expanded_ver18_cpu') - backend_test.exclude(r'test_softsign_example_expanded_ver18_cpu') - backend_test.exclude(r'test_softsign_expanded_ver18_cpu') def disabled_tests_onnx_1_14_0(backend_test): From f5411e07b36056c335218c9f1a6d6a973d04a3b7 Mon Sep 17 00:00:00 2001 From: Artur Wojcik Date: Wed, 4 Oct 2023 21:46:51 +0200 Subject: [PATCH 12/19] fix call to std::isfinite() on Windows (#2283) --- src/cpp_generator.cpp | 4 ++-- src/include/migraphx/op/convert.hpp | 2 +- src/include/migraphx/op/isnan.hpp | 2 +- src/include/migraphx/verify.hpp | 6 ++---- src/program.cpp | 2 +- 5 files changed, 7 insertions(+), 9 deletions(-) diff --git a/src/cpp_generator.cpp b/src/cpp_generator.cpp index 464570b82a6..dedcea6b7f7 100644 --- a/src/cpp_generator.cpp +++ b/src/cpp_generator.cpp @@ -213,13 +213,13 @@ cpp_generator::function cpp_generator::generate_module(const module& m, ins->get_literal().visit([&](auto v) { assert(v.size() == 1); auto x = v.front(); - if(std::isinf(x)) + if(std::isinf(static_cast(x))) { string_literal = "__builtin_huge_val()"; if(x < 0) string_literal = "-__builtin_huge_val()"; } - else if(std::isnan(x)) + else if(std::isnan(static_cast(x))) string_literal = "__builtin_nan()"; else string_literal = ins->get_literal().to_string(); diff --git a/src/include/migraphx/op/convert.hpp b/src/include/migraphx/op/convert.hpp index 76874274a51..4a52348049d 100644 --- a/src/include/migraphx/op/convert.hpp +++ b/src/include/migraphx/op/convert.hpp @@ -68,7 +68,7 @@ struct convert : unary auto y = x; shape::visit(type, [&](auto as) { // clamping value between target_type's max and min doesn't work for NaNs, - if(std::isnan(x)) + if(std::isnan(static_cast(x))) { y = as.nan(); } diff --git a/src/include/migraphx/op/isnan.hpp b/src/include/migraphx/op/isnan.hpp index d953cb1b137..130efc53a78 100644 --- a/src/include/migraphx/op/isnan.hpp +++ b/src/include/migraphx/op/isnan.hpp @@ -35,7 +35,7 @@ struct isnan : unary { auto apply() const { - return [](auto x) { return std::isnan(x); }; + return [](auto x) { return std::isnan(static_cast(x)); }; } std::string name() const { return "isnan"; } diff --git a/src/include/migraphx/verify.hpp b/src/include/migraphx/verify.hpp index fbea292238b..33038163fa8 100644 --- a/src/include/migraphx/verify.hpp +++ b/src/include/migraphx/verify.hpp @@ -90,8 +90,7 @@ struct not_finite_fn template bool operator()(T x) const { - using std::isfinite; - return not isfinite(x); + return not std::isfinite(static_cast(x)); } }; static constexpr not_finite_fn not_finite{}; @@ -101,8 +100,7 @@ struct compare_mag_fn template bool operator()(T x, U y) const { - using std::fabs; - return fabs(x) < fabs(y); + return std::fabs(x) < std::fabs(y); } }; static constexpr compare_mag_fn compare_mag{}; diff --git a/src/program.cpp b/src/program.cpp index 7e0dbb8b7a9..4efe90ed8ed 100644 --- a/src/program.cpp +++ b/src/program.cpp @@ -347,7 +347,7 @@ void program::finalize() template std::string classify(T x) { - switch(std::fpclassify(x)) + switch(std::fpclassify(static_cast(x))) { case FP_INFINITE: return "inf"; case FP_NAN: return "nan"; From 65c37c3d99c6f8fdef6b466a8d5718ee015f85c7 Mon Sep 17 00:00:00 2001 From: Ahsan Saghir <142340507+ahsan-ca@users.noreply.github.com> Date: Thu, 5 Oct 2023 09:17:46 -0400 Subject: [PATCH 13/19] Specify full path for libamdhip64.so (#2264) --- test/py/test_gpu_async.py | 49 +++++++++++++++++++++++++++++++++++++-- 1 file changed, 47 insertions(+), 2 deletions(-) diff --git a/test/py/test_gpu_async.py b/test/py/test_gpu_async.py index 92cf9249efe..728c67490e2 100644 --- a/test/py/test_gpu_async.py +++ b/test/py/test_gpu_async.py @@ -1,7 +1,7 @@ ##################################################################################### # The MIT License (MIT) # -# Copyright (c) 2015-2022 Advanced Micro Devices, Inc. All rights reserved. +# Copyright (c) 2015-2023 Advanced Micro Devices, Inc. All rights reserved. # # Permission is hereby granted, free of charge, to any person obtaining a copy # of this software and associated documentation files (the "Software"), to deal @@ -23,10 +23,55 @@ ##################################################################################### import migraphx import ctypes +import os +import glob def test_conv_relu(): - hip = ctypes.cdll.LoadLibrary("libamdhip64.so") + # Full path of the library is needed to fix an issue on sles + # where the library is not loaded otherwise. + # We check for the presence of library at the following paths, + # in the order listed below: + # + # 1. 'rocm_path' environment variable + # 2. /opt/rocm + # 3. /opt/rocm-* + # + # If the library is not found at any of these paths, we fall back + # to the library path being detected automatically. + + library = "libamdhip64.so" + + # Environment variable containing path to rocm + rocm_path_env_var = "rocm_path" + + # Check for rocm_path, default to /opt/rocm if it does not exist. + rocm_path_var = os.getenv(rocm_path_env_var, default="/opt/rocm") + + # Join the paths to the library to get full path, + # e.g. /opt/rocm/lib/libamdhip64.so + library_file = os.path.join(rocm_path_var, "lib", library) + + # Check if the library file exists at the specified path + if os.path.exists(library_file): + # Replace library name by full path to the library + library = library_file + else: + # Pattern match to look for path to different + # rocm versions: /opt/rocm-* + rocm_path_pattern = "/opt/rocm-*/lib/libamdhip64.so" + matching_libraries = glob.glob(rocm_path_pattern) + + if matching_libraries: + # Replace library name by full path to the first + # library found. + library = matching_libraries[0] + + # Loads library either by using the full path to the + # library, if it has been detected earlier, + # or, proceeds to load the library based on the name + # of the library. + hip = ctypes.cdll.LoadLibrary(library) p = migraphx.parse_onnx("conv_relu_maxpool_test.onnx") print(p) From c7f0fbc4418b3ed300311580954ad9c2400ab898 Mon Sep 17 00:00:00 2001 From: Charlie Lin Date: Fri, 6 Oct 2023 08:20:47 -0400 Subject: [PATCH 14/19] compile options can be set for MIGX (#2298) --- tools/accuracy/accuracy_checker.py | 28 +++++++++++++++++++++++++++- 1 file changed, 27 insertions(+), 1 deletion(-) diff --git a/tools/accuracy/accuracy_checker.py b/tools/accuracy/accuracy_checker.py index f5db40b8652..d368ca2a29e 100644 --- a/tools/accuracy/accuracy_checker.py +++ b/tools/accuracy/accuracy_checker.py @@ -82,6 +82,27 @@ def parse_args(): default=False, help='Turn on ort VERBOSE logging via session options') + parser.add_argument( + '--disable-offload-copy', + dest="offload_copy", + action='store_false', + default=True, + help= + 'Disable offload copying (user must handle copy to and from device)') + + parser.add_argument( + '--disable-fast-math', + dest="fast_math", + action='store_false', + default=True, + help='Disable fast math optimizations (etc: rewrite_gelu)') + + parser.add_argument('--exhaustive_tune', + dest="exhaustive_tune", + action='store_true', + default=False, + help='Enable exhaustive tuning for solutions') + args = parser.parse_args() return args @@ -177,7 +198,12 @@ def main(): print(model) if not args.ort_run: - model.compile(migraphx.get_target(args.target)) + model.compile( + migraphx.get_target(args.target), + offload_copy=args.offload_copy, + fast_math=args.fast_math, + exhaustive_tune=args.exhaustive_tune, + ) params = {} test_inputs = {} From 9d8331b4e2595fe7cedaa44c4a46a0c86addf2ef Mon Sep 17 00:00:00 2001 From: Artur Wojcik Date: Fri, 6 Oct 2023 21:13:32 +0200 Subject: [PATCH 15/19] prepare for Windows resources with resource script files (#1999) --- cmake/Embed.cmake | 29 ++++++++++--------- src/compile_src.cpp | 2 +- src/include/migraphx/compile_src.hpp | 14 +++++++-- src/targets/gpu/CMakeLists.txt | 10 ++++++- src/targets/gpu/compile_hip.cpp | 8 ++--- src/targets/gpu/compile_hip_code_object.cpp | 22 ++++++-------- .../gpu/include/migraphx/gpu/compile_hip.hpp | 5 +--- src/targets/gpu/jit/ck_gemm.cpp | 3 +- test/gpu/jit.cpp | 2 +- test/gpu/stream_sync.cpp | 2 +- test/jit.cpp | 4 +-- test/multi_target/multitarget_test.cpp | 1 - 12 files changed, 54 insertions(+), 48 deletions(-) diff --git a/cmake/Embed.cmake b/cmake/Embed.cmake index 5a120e18339..f3ef595c995 100755 --- a/cmake/Embed.cmake +++ b/cmake/Embed.cmake @@ -77,16 +77,17 @@ function(generate_embed_source EMBED_NAME) list(GET PARSE_FILES ${idx} FILE) set(START_SYMBOL "_binary_${SYMBOL}_start") - set(END_SYMBOL "_binary_${SYMBOL}_end") + set(LENGTH_SYMBOL "_binary_${SYMBOL}_length") if(EMBED_USE_LD) string(APPEND EXTERNS " - extern const char ${START_SYMBOL}[]; - extern const char ${END_SYMBOL}[]; +extern const char ${START_SYMBOL}[]; +extern const size_t _binary_${SYMBOL}_size; +const auto ${LENGTH_SYMBOL} = reinterpret_cast(&_binary_${SYMBOL}_size); ") else() string(APPEND EXTERNS " - extern const char ${START_SYMBOL}[]; - extern const char* ${END_SYMBOL}; +extern const char ${START_SYMBOL}[]; +extern const size_t ${LENGTH_SYMBOL}; ") endif() @@ -97,23 +98,22 @@ function(generate_embed_source EMBED_NAME) endif() string(APPEND INIT_KERNELS " - { \"${BASE_NAME}\", { ${START_SYMBOL}, ${END_SYMBOL}} }, - ") + { \"${BASE_NAME}\", { ${START_SYMBOL}, ${LENGTH_SYMBOL}} },") endforeach() file(WRITE "${PARSE_HEADER}" " +#include #include -#include #include -const std::unordered_map>& ${EMBED_NAME}(); +std::unordered_map ${EMBED_NAME}(); ") file(WRITE "${PARSE_SRC}" " #include <${EMBED_NAME}.hpp> ${EXTERNS} -const std::unordered_map>& ${EMBED_NAME}() +std::unordered_map ${EMBED_NAME}() { - static const std::unordered_map> result = {${INIT_KERNELS}}; + static std::unordered_map result = {${INIT_KERNELS}}; return result; } ") @@ -154,9 +154,10 @@ function(embed_file OUTPUT_FILE OUTPUT_SYMBOL FILE) # removes trailing comma string(REGEX REPLACE ", $" "" ARRAY_VALUES ${ARRAY_VALUES}) file(WRITE "${OUT_FILE}" " - extern const char _binary_${SYMBOL}_start[] = { ${ARRAY_VALUES} }; - extern const char* _binary_${SYMBOL}_end = _binary_${SYMBOL}_start + sizeof(_binary_${SYMBOL}_start); - \n") +#include +extern const char _binary_${SYMBOL}_start[] = { ${ARRAY_VALUES} }; +extern const size_t _binary_${SYMBOL}_length = sizeof(_binary_${SYMBOL}_start); +") endif() endforeach() endfunction() diff --git a/src/compile_src.cpp b/src/compile_src.cpp index 6466dad8493..2029f734519 100644 --- a/src/compile_src.cpp +++ b/src/compile_src.cpp @@ -46,7 +46,7 @@ std::vector src_compiler::compile(const std::vector& srcs) const fs::path full_path = td.path / src.path; fs::path parent_path = full_path.parent_path(); fs::create_directories(parent_path); - write_buffer(full_path.string(), src.content.first, src.len()); + write_buffer(full_path.string(), src.content.data(), src.content.size()); if(src.path.extension().string() == ".cpp") { params += " " + src.path.filename().string(); diff --git a/src/include/migraphx/compile_src.hpp b/src/include/migraphx/compile_src.hpp index 6803a66af65..9baf2bc6cba 100644 --- a/src/include/migraphx/compile_src.hpp +++ b/src/include/migraphx/compile_src.hpp @@ -37,8 +37,18 @@ inline namespace MIGRAPHX_INLINE_NS { struct src_file { fs::path path; - std::pair content; - std::size_t len() const { return content.second - content.first; } + std::string_view content; + + src_file() = default; + src_file(fs::path file_path, std::string_view file_content) + : path{std::move(file_path)}, content{file_content} + { + } + + explicit src_file(const std::pair& pair) + : path{pair.first}, content{pair.second} + { + } }; struct MIGRAPHX_EXPORT src_compiler diff --git a/src/targets/gpu/CMakeLists.txt b/src/targets/gpu/CMakeLists.txt index d6dcd182f8e..c53860552b7 100644 --- a/src/targets/gpu/CMakeLists.txt +++ b/src/targets/gpu/CMakeLists.txt @@ -48,10 +48,18 @@ else() set(MIGRAPHX_USE_HIPRTC ON CACHE BOOL "Use hipRTC APIs") endif() -include(Embed) file(GLOB KERNEL_FILES CONFIGURE_DEPENDS ${CMAKE_CURRENT_SOURCE_DIR}/kernels/include/migraphx/kernels/*.hpp) message(STATUS "KERNEL_FILES: ${KERNEL_FILES}") + +if(WIN32) + # TODO: re-enable when CK is ported to Windows + list(REMOVE_ITEM KERNEL_FILES + ${CMAKE_CURRENT_SOURCE_DIR}/kernels/include/migraphx/kernels/ck_gemm.hpp + ${CMAKE_CURRENT_SOURCE_DIR}/kernels/include/migraphx/kernels/ck.hpp) +endif() + +include(Embed) add_embed_library(migraphx_kernels ${KERNEL_FILES} RELATIVE ${CMAKE_CURRENT_SOURCE_DIR}/kernels/include/) configure_file(device/targets.hpp.in include/migraphx/gpu/device/targets.hpp) diff --git a/src/targets/gpu/compile_hip.cpp b/src/targets/gpu/compile_hip.cpp index 30565fadfe6..14e64c1e97c 100644 --- a/src/targets/gpu/compile_hip.cpp +++ b/src/targets/gpu/compile_hip.cpp @@ -248,7 +248,7 @@ compile_hip_src(const std::vector& srcs, std::string params, const std { if(src.path.extension() != ".cpp") continue; - std::cout << std::string(src.content.first, src.len()) << std::endl; + std::cout << std::string(src.content) << std::endl; } } auto p = dynamic_loader::path(&compile_hip_src_with_hiprtc); @@ -338,7 +338,7 @@ compile_hip_src(const std::vector& srcs, std::string params, const std { if(src.path.extension() != ".cpp") continue; - std::cout << std::string(src.content.first, src.len()) << std::endl; + std::cout << std::string(src.content) << std::endl; } } @@ -359,9 +359,7 @@ bool hip_has_flags(const std::vector& flags) join_strings(flags, " ") + " -x hip -c --offload-arch=gfx900 --cuda-device-only"; std::string src; - src_file input; - input.path = "main.cpp"; - input.content = std::make_pair(src.data(), src.data() + src.size()); + src_file input{"main.cpp", src}; try { diff --git a/src/targets/gpu/compile_hip_code_object.cpp b/src/targets/gpu/compile_hip_code_object.cpp index 828093e96b3..74aa9d24138 100644 --- a/src/targets/gpu/compile_hip_code_object.cpp +++ b/src/targets/gpu/compile_hip_code_object.cpp @@ -172,21 +172,17 @@ operation compile_hip_code_object(const std::string& content, hip_compile_option assert(options.inputs.size() == options.virtual_inputs.size() or options.virtual_inputs.empty()); std::vector srcs = options.additional_src_files; - std::transform(migraphx_kernels().begin(), - migraphx_kernels().end(), - std::back_inserter(srcs), - [](auto&& p) { - auto&& name = p.first; - auto&& c = p.second; - auto path = name; - return src_file{path, c}; - }); - srcs.push_back(src_file{fs::path{"main.cpp"}, - std::make_pair(content.data(), content.data() + content.size())}); + static auto kernels{::migraphx_kernels()}; + std::transform( + kernels.begin(), + kernels.end(), + std::back_inserter(srcs), + [](const std::pair& elem) { return src_file{elem}; }); + srcs.emplace_back("main.cpp", content); auto args_hpp = generate_args_hpp(options.virtual_inputs.empty() ? options.inputs : options.virtual_inputs); - srcs.push_back(src_file{fs::path{"args.hpp"}, - std::make_pair(args_hpp.data(), args_hpp.data() + args_hpp.size())}); + srcs.emplace_back("args.hpp", args_hpp); + options.params += " -DMIGRAPHX_NGLOBAL=" + std::to_string(options.global); options.params += " -DMIGRAPHX_NLOCAL=" + std::to_string(options.local); options.params += " " + join_strings(compiler_warnings(), " "); diff --git a/src/targets/gpu/include/migraphx/gpu/compile_hip.hpp b/src/targets/gpu/include/migraphx/gpu/compile_hip.hpp index 3607cd46a40..447cdab126d 100644 --- a/src/targets/gpu/include/migraphx/gpu/compile_hip.hpp +++ b/src/targets/gpu/include/migraphx/gpu/compile_hip.hpp @@ -45,10 +45,7 @@ MIGRAPHX_DECLARE_ENV_VAR(MIGRAPHX_ENABLE_HIPRTC_WORKAROUNDS); struct hiprtc_src_file { hiprtc_src_file() = default; - hiprtc_src_file(const src_file& s) - : path(s.path.string()), content(s.content.first, s.content.second) - { - } + hiprtc_src_file(const src_file& s) : path(s.path.string()), content(s.content) {} std::string path; std::string content; template diff --git a/src/targets/gpu/jit/ck_gemm.cpp b/src/targets/gpu/jit/ck_gemm.cpp index 65bed54a800..2937f653c09 100644 --- a/src/targets/gpu/jit/ck_gemm.cpp +++ b/src/targets/gpu/jit/ck_gemm.cpp @@ -112,8 +112,7 @@ static std::vector create_ck_headers() std::vector srcs; std::transform( header_strings.begin(), header_strings.end(), std::back_inserter(srcs), [&](auto&& p) { - return src_file{fs::path{p.first}, - {p.second.data(), p.second.data() + p.second.size()}}; + return src_file{p.first, p.second}; }); return srcs; } diff --git a/test/gpu/jit.cpp b/test/gpu/jit.cpp index 90c19c7c490..b92f1419310 100644 --- a/test/gpu/jit.cpp +++ b/test/gpu/jit.cpp @@ -155,7 +155,7 @@ int main() {} migraphx::src_file make_src_file(const std::string& name, const std::string& content) { - return {name, std::make_pair(content.data(), content.data() + content.size())}; + return {name, content}; } TEST_CASE(simple_compile_hip) diff --git a/test/gpu/stream_sync.cpp b/test/gpu/stream_sync.cpp index da7c00e5a0a..e9d2c97d0b6 100644 --- a/test/gpu/stream_sync.cpp +++ b/test/gpu/stream_sync.cpp @@ -64,7 +64,7 @@ int main() {} migraphx::src_file make_src_file(const std::string& name, const std::string& content) { - return {name, std::make_pair(content.data(), content.data() + content.size())}; + return {name, content}; } hip_stream_ptr get_stream() diff --git a/test/jit.cpp b/test/jit.cpp index a1e5253846e..c269702d6a4 100644 --- a/test/jit.cpp +++ b/test/jit.cpp @@ -48,9 +48,7 @@ compile_function(const std::string& src, const std::string& flags, const std::st migraphx::src_compiler compiler; compiler.flags = flags + "-std=c++14 -fPIC -shared"; compiler.output = "libsimple.so"; - migraphx::src_file f; - f.path = "main.cpp"; - f.content = std::make_pair(src.data(), src.data() + src.size()); + migraphx::src_file f{"main.cpp", src}; auto image = compiler.compile({f}); return migraphx::dynamic_loader{image}.get_function(fname); } diff --git a/test/multi_target/multitarget_test.cpp b/test/multi_target/multitarget_test.cpp index 6b25405ac87..c375612d760 100644 --- a/test/multi_target/multitarget_test.cpp +++ b/test/multi_target/multitarget_test.cpp @@ -37,7 +37,6 @@ #include #include #include -#include #include #include #include From 24146af67cc18655bc3d5446a0b43746749fd912 Mon Sep 17 00:00:00 2001 From: Chris Austen Date: Fri, 6 Oct 2023 17:18:50 -0400 Subject: [PATCH 16/19] resolve ci hang with mlir (#2290) --- requirements.txt | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/requirements.txt b/requirements.txt index 6e24e0110a4..3a06a254528 100755 --- a/requirements.txt +++ b/requirements.txt @@ -29,4 +29,4 @@ pybind/pybind11@d159a563383d10c821ba7b2a71905d1207db6de4 --build msgpack/msgpack-c@cpp-3.3.0 -DMSGPACK_BUILD_TESTS=Off sqlite3@3.17 -DCMAKE_POSITION_INDEPENDENT_CODE=On ROCmSoftwarePlatform/composable_kernel@a22e479b8e1557961039db2d5c5ff89cff35e86b -DCK_BUILD_JIT_LIB=On -DCMAKE_POSITION_INDEPENDENT_CODE=On -ROCmSoftwarePlatform/rocMLIR@a48dfb1f163fb0b38369e73e580968b72e85b594 -DBUILD_FAT_LIBROCKCOMPILER=On +ROCmSoftwarePlatform/rocMLIR@12748a3402c069f733ea7f2ba1f8d8a070b3622a -DBUILD_FAT_LIBROCKCOMPILER=On From 9316c0076b62c4ba3baf90c0c291c3bea648488e Mon Sep 17 00:00:00 2001 From: Charlie Lin Date: Fri, 6 Oct 2023 17:24:58 -0400 Subject: [PATCH 17/19] Add explanatory documentation to `allocate` (#2301) --- src/include/migraphx/op/allocate.hpp | 13 +++++++++++++ 1 file changed, 13 insertions(+) diff --git a/src/include/migraphx/op/allocate.hpp b/src/include/migraphx/op/allocate.hpp index 5d1ca929ffb..33ea6bb2260 100644 --- a/src/include/migraphx/op/allocate.hpp +++ b/src/include/migraphx/op/allocate.hpp @@ -33,6 +33,19 @@ namespace migraphx { inline namespace MIGRAPHX_INLINE_NS { namespace op { +/** + * Static allocate: + * No inputs: `allocate()` + * `this.s` attribute set to the static output shape of the buffer. + * + * Dynamic allocate: + * One input: `allocate(output_dims)` + * `output_dims` are the output buffer dimensions and has a static shape. + * Either `this.s` or `this.buf_type` must be set to calculate the dynamic output shape at compute + * time. If `this.buf_type` is set, the compute_shape() of allocate at compile time will have + * dynamic_dimensions from {0, max_int} with rank = output_dims.ndim(). If `this.s` is set then the + * compute_shape() will output `this.s`; `this.s` should be a dynamic shape. + */ struct allocate { shape s{}; From f3f65985cc522dab71deccea5db02370acf575e4 Mon Sep 17 00:00:00 2001 From: tvukovic-amd <127323445+tvukovic-amd@users.noreply.github.com> Date: Sat, 7 Oct 2023 00:30:18 +0200 Subject: [PATCH 18/19] Added fix for test_msgpack_float and test_msgpack_object (#2155) --- test/msgpack.cpp | 19 +++++++++++-------- 1 file changed, 11 insertions(+), 8 deletions(-) diff --git a/test/msgpack.cpp b/test/msgpack.cpp index 8b4e8ce1d42..82912aeac26 100644 --- a/test/msgpack.cpp +++ b/test/msgpack.cpp @@ -97,9 +97,12 @@ TEST_CASE(test_msgpack_bool) TEST_CASE(test_msgpack_float) { - migraphx::value v = 3.0; + // changed all double values in this code to not end with .0 because on msgpack for Windows if + // input type is double and ends with .0 it could be converted to uint64_t or int64_t and the + // goal of these functions is to test double without conversions + migraphx::value v = 3.01; auto buffer = migraphx::to_msgpack(v); - EXPECT(buffer == msgpack_buffer(3.0)); + EXPECT(buffer == msgpack_buffer(3.01)); EXPECT(migraphx::from_msgpack(buffer) == v); } @@ -129,10 +132,10 @@ TEST_CASE(test_msgpack_empty_array) TEST_CASE(test_msgpack_object) { - migraphx::value v = {{"one", 1.0}, {"three", 3.0}, {"two", 2.0}}; + migraphx::value v = {{"one", 1.01}, {"three", 3.01}, {"two", 2.01}}; auto buffer = migraphx::to_msgpack(v); EXPECT(buffer == msgpack_buffer(std::map{ - {"one", 1.0}, {"three", 3.0}, {"two", 2.0}})); + {"one", 1.01}, {"three", 3.01}, {"two", 2.01}})); EXPECT(migraphx::from_msgpack(buffer) == v); } @@ -157,17 +160,17 @@ struct foo TEST_CASE(test_msgpack_object_class) { - migraphx::value v = {{"a", 1.0}, {"b", "abc"}}; + migraphx::value v = {{"a", 1.01}, {"b", "abc"}}; auto buffer = migraphx::to_msgpack(v); - EXPECT(buffer == msgpack_buffer(foo{1.0, "abc"})); + EXPECT(buffer == msgpack_buffer(foo{1.01, "abc"})); EXPECT(migraphx::from_msgpack(buffer) == v); } TEST_CASE(test_msgpack_array_class) { - migraphx::value v = {{{"a", 1.0}, {"b", "abc"}}, {{"a", 3.0}, {"b", "xyz"}}}; + migraphx::value v = {{{"a", 1.01}, {"b", "abc"}}, {{"a", 3.01}, {"b", "xyz"}}}; auto buffer = migraphx::to_msgpack(v); - EXPECT(buffer == msgpack_buffer(std::vector{foo{1.0, "abc"}, foo{3.0, "xyz"}})); + EXPECT(buffer == msgpack_buffer(std::vector{foo{1.01, "abc"}, foo{3.01, "xyz"}})); EXPECT(migraphx::from_msgpack(buffer) == v); } From 1082f667f6b771244a123726e0371577d749af72 Mon Sep 17 00:00:00 2001 From: Artur Wojcik Date: Sat, 7 Oct 2023 00:30:49 +0200 Subject: [PATCH 19/19] add missing DLL symbols exports (#2281) --- src/include/migraphx/normalize_attributes.hpp | 2 ++ src/include/migraphx/pad_calc.hpp | 1 + src/targets/gpu/device/targets.hpp.in | 6 +++++- 3 files changed, 8 insertions(+), 1 deletion(-) diff --git a/src/include/migraphx/normalize_attributes.hpp b/src/include/migraphx/normalize_attributes.hpp index e88003e6d85..61887af3f59 100644 --- a/src/include/migraphx/normalize_attributes.hpp +++ b/src/include/migraphx/normalize_attributes.hpp @@ -52,6 +52,7 @@ using dependent_type = typename select_dependent_type::type; * \param attr_val the normalize_axes attributes from the operator * \param prefix error message prefix */ +MIGRAPHX_EXPORT std::vector normalize_axes(const std::vector& axes, const shape& input_shape, const value& attr_val, @@ -67,6 +68,7 @@ std::vector normalize_axes(const std::vector& axes, * \param attr_val the normalize_axes attributes from the operator * \param prefix error message prefix */ +MIGRAPHX_EXPORT std::vector normalize_indices(const std::vector& indices, const std::vector& axes, const shape& input_shape, diff --git a/src/include/migraphx/pad_calc.hpp b/src/include/migraphx/pad_calc.hpp index a17c0bc3028..cb5972fb644 100644 --- a/src/include/migraphx/pad_calc.hpp +++ b/src/include/migraphx/pad_calc.hpp @@ -64,6 +64,7 @@ shape compute_padded_shape(const shape& input, // Used for dynamic auto padding of pooling operators where padding needs to be computed at // evaulation time. +MIGRAPHX_EXPORT shape compute_padded_pool_shape(const shape& input, const shape& kernel, const std::vector& padding, diff --git a/src/targets/gpu/device/targets.hpp.in b/src/targets/gpu/device/targets.hpp.in index 0d73f7be36d..0a0e19aba6b 100644 --- a/src/targets/gpu/device/targets.hpp.in +++ b/src/targets/gpu/device/targets.hpp.in @@ -24,7 +24,7 @@ #ifndef MIGRAPHX_GUARD_DEVICE_TARGETS_CPP #define MIGRAPHX_GUARD_DEVICE_TARGETS_CPP -#include +#include #include #include @@ -34,9 +34,13 @@ namespace gpu { namespace device { #define MIGRAPHX_GPU_TARGETS "@GPU_TARGETS@" // NOLINT +MIGRAPHX_DEVICE_EXPORT const std::vector& get_targets(); + +MIGRAPHX_DEVICE_EXPORT std::string get_targets_as_string(); +MIGRAPHX_DEVICE_EXPORT std::string get_device_name(); } // namespace device