Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merge Navi changes for int8 unit tests and verification mlir issues #2393

Merged
merged 6 commits into from
Nov 10, 2023
Merged
Show file tree
Hide file tree
Changes from 3 commits
Commits
File filter

Filter by extension

Filter by extension


Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
43 changes: 43 additions & 0 deletions CHANGELOG.md
Original file line number Diff line number Diff line change
Expand Up @@ -2,6 +2,49 @@

Full documentation for MIGraphX is available at [MIGraphX Documentation](https://rocmdocs.amd.com/projects/AMDMIGraphX/en/latest/).

## MIGraphX 2.8 for ROCm 6.0.0
### Added
- Support for MI300 GPUs
- Support for TorchMIGraphX via PyTorch
- Boosted overall performance by integrating rocMLIR
- INT8 support for ONNX Runtime
- Support for ONNX version 1.14.1
- Added operators Qlinearadd, QlinearGlobalAveragePool, Qlinearconv, Shrink, CastLike, and RandomUniform operators
- Added an error message when gpu_targets is not set when compiling migraphx
- Added parameter to set tolerances with migraphx-driver verify
- Added support for MXR files >4 GB
- Added MIGRAPHX_TRACE_MLIR flag
- BETA added capability to use ROCm Composable Kernels via environment variable MIGRAPHX_ENABLE_CK=1

### Optimizations
- Improved performance support for INT8
- Improved time percision while benchmarking candidate kernels from CK or MLIR
- Remove contiguous from reshape parsing
- Updated ConstantOfShape operator to support Dynamic Batch
- Simplifies dynamic shapes related operators to their static versions if possible
- Improved debugging tools for accuracy issues
- Print warning about miopen_fusion while generating mxr
- General reduction in system memory usage during model compilation
- Created additional fusion opportunities during model compilation
- Improved debugging for matchers
- Improved general debug messages

### Fixed
- Fixed scatter operator for nonstandard shapes with some models from ONNX Model Zoo
- Provided a compile option to improve accuracy of some models by disabling Fast-Math
- Improved layernorm + pointwise fusion matching to ignore arguments order
- Fixed accuracy issue with ROIAlign operator
- Fixed Trilu operator computation logic
- Fixed support for the DETR model

### Changed
- Changed migraphx version to 2.8
- Extracted test packages as its own separate deb file when building migraphx from source

### Removed
- Removed building Python 2.7 bindings


## MIGraphX 2.7 for ROCm 5.7.0
### Added
- Enabled hipRTC to not require dev packages for migraphx runtime and allow the ROCm install to be in a different directory than it was during build time
Expand Down
2 changes: 1 addition & 1 deletion requirements.txt
Original file line number Diff line number Diff line change
Expand Up @@ -29,4 +29,4 @@ pybind/pybind11@d159a563383d10c821ba7b2a71905d1207db6de4 --build
msgpack/[email protected] -DMSGPACK_BUILD_TESTS=Off
[email protected] -DCMAKE_POSITION_INDEPENDENT_CODE=On
ROCmSoftwarePlatform/composable_kernel@70eefcf4f263aa5c25f3c9ff0db8f6f199ef0fb9 -DCK_BUILD_JIT_LIB=On -DCMAKE_POSITION_INDEPENDENT_CODE=On
ROCmSoftwarePlatform/rocMLIR@507bb94ce7873786486d296ec81d2eadaab49003 -DBUILD_FAT_LIBROCKCOMPILER=On
ROCmSoftwarePlatform/rocMLIR@3700afd2564e21267a4d1fd8f1f80465f45daa93 -DBUILD_FAT_LIBROCKCOMPILER=On
9 changes: 6 additions & 3 deletions src/driver/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -55,6 +55,9 @@ if(MIGRAPHX_ENABLE_PYTHON)
target_compile_definitions(driver PRIVATE MIGRAPHX_ENABLE_PYTHON)
endif()

rocm_install_targets(
TARGETS driver
)
# For ASAN Enabled Build no need to package binaries
if( NOT ENABLE_ASAN_PACKAGING )
rocm_install_targets(
TARGETS driver
)
endif()
2 changes: 1 addition & 1 deletion src/onnx/parse_slice.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -137,7 +137,7 @@ struct parse_slice : op_parser<parse_slice>
sd.always_insert(args.at(0));

// If axes arg is not given, the default is all of them.
if(sd.op.axes.empty() and sd.op_args.size() < 3)
if(sd.op.axes.empty() and sd.op_args.size() <= 3)
{
std::vector<int64_t> axes(args[0]->get_shape().ndim());
std::iota(axes.begin(), axes.end(), int64_t{0});
Expand Down
2 changes: 1 addition & 1 deletion src/program.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -936,7 +936,7 @@ void program::perf_report(std::ostream& os,
os << std::endl;

os << "Batch size: " << batch << std::endl;
os << "Rate: " << rate * batch << "/sec" << std::endl;
os << "Rate: " << rate * batch << " inferences/sec" << std::endl;
os << "Total time: " << total_time << "ms" << std::endl;
os << "Total instructions time: " << total_instruction_time << "ms" << std::endl;
os << "Overhead time: " << overhead_time << "ms"
Expand Down
2 changes: 1 addition & 1 deletion src/quantization.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -147,8 +147,8 @@ void quantize_int8(program& prog,

run_passes(prog,
{quantize_int8_pass{ins_names, *int8_quant_params},
optimize_module{},
simplify_qdq{},
optimize_module{},
dead_code_elimination{}});
}

Expand Down
3 changes: 2 additions & 1 deletion src/targets/gpu/fuse_mlir.cpp
Original file line number Diff line number Diff line change
Expand Up @@ -361,7 +361,8 @@ struct find_mlir_standalone_op
return;

static size_t counter = 0;
module_ref mm = mpm.create_module("mlir_" + std::to_string(counter++));
module_ref mm =
mpm.create_module("mlir_" + conv_based_op->name() + std::to_string(counter++));
mm->set_bypass();
auto [anchor_op, top_inputs] = fuse_input_ops_and_gemm_based_op(mm, conv_based_op);
mm->add_return({anchor_op});
Expand Down
10 changes: 7 additions & 3 deletions src/targets/gpu/hiprtc/CMakeLists.txt
Original file line number Diff line number Diff line change
Expand Up @@ -28,6 +28,10 @@ add_executable(migraphx-hiprtc-driver
rocm_clang_tidy_check(migraphx-hiprtc-driver)
target_link_libraries(migraphx-hiprtc-driver PRIVATE migraphx_gpu)
add_dependencies(migraphx_all_targets migraphx-hiprtc-driver)
rocm_install_targets(
TARGETS migraphx-hiprtc-driver
)

# For ASAN Enabled Build no need to package binaries
if( NOT ENABLE_ASAN_PACKAGING )
rocm_install_targets(
TARGETS migraphx-hiprtc-driver
)
endif()
2 changes: 1 addition & 1 deletion test/py/requirements.txt
Original file line number Diff line number Diff line change
Expand Up @@ -22,4 +22,4 @@
# THE SOFTWARE.
#####################################################################################

numpy==1.21.6
numpy==1.19.5
Loading