mirror of
https://github.com/saymrwulf/onnxruntime.git
synced 2026-07-29 20:14:01 +00:00
Revert "[MIGraphX EP] enable compilation and execution on Windows (21084)" (#21132)
### Description
This reverts commit 1d7bf56947 because it
broken the AMD GPU CI pipeline. Sorry when I reviewed the PR I forgot to
run the AMD GPU CI pipeline.
Will revert the PR first then ask the author to fix the issue.
This commit is contained in:
Changming Sun
GitHub
parent
69d522f4e9
commit
f5625b8858
22 changed files with 173 additions and 430 deletions
|
|
@ -1472,6 +1472,9 @@ if (onnxruntime_USE_CUDA)
|
|||
endif()
|
||||
|
||||
if (onnxruntime_USE_MIGRAPHX)
|
||||
if (WIN32)
|
||||
message(FATAL_ERROR "MIGraphX does not support build in Windows!")
|
||||
endif()
|
||||
set(AMD_MIGRAPHX_HOME ${onnxruntime_MIGRAPHX_HOME})
|
||||
endif()
|
||||
|
||||
|
|
|
|||
|
|
@ -19,25 +19,23 @@
|
|||
endif()
|
||||
|
||||
# Add search paths for default rocm installation
|
||||
list(APPEND CMAKE_PREFIX_PATH /opt/rocm/hcc /opt/rocm/hip /opt/rocm $ENV{HIP_PATH})
|
||||
list(APPEND CMAKE_PREFIX_PATH /opt/rocm/hcc /opt/rocm/hip /opt/rocm)
|
||||
|
||||
# Suppress the warning about the small capitals of the package name
|
||||
cmake_policy(SET CMP0144 NEW)
|
||||
find_package(hip)
|
||||
find_package(migraphx PATHS ${AMD_MIGRAPHX_HOME})
|
||||
|
||||
if(WIN32 AND NOT HIP_PLATFORM)
|
||||
set(HIP_PLATFORM "amd")
|
||||
endif()
|
||||
find_package(miopen)
|
||||
find_package(rocblas)
|
||||
|
||||
find_package(hip REQUIRED)
|
||||
find_package(migraphx REQUIRED PATHS ${AMD_MIGRAPHX_HOME})
|
||||
|
||||
set(migraphx_libs migraphx::c hip::host)
|
||||
set(migraphx_libs migraphx::c hip::host MIOpen roc::rocblas)
|
||||
|
||||
file(GLOB_RECURSE onnxruntime_providers_migraphx_cc_srcs CONFIGURE_DEPENDS
|
||||
"${ONNXRUNTIME_ROOT}/core/providers/migraphx/*.h"
|
||||
"${ONNXRUNTIME_ROOT}/core/providers/migraphx/*.cc"
|
||||
"${ONNXRUNTIME_ROOT}/core/providers/shared_library/*.h"
|
||||
"${ONNXRUNTIME_ROOT}/core/providers/shared_library/*.cc"
|
||||
"${ONNXRUNTIME_ROOT}/core/providers/rocm/rocm_stream_handle.h"
|
||||
"${ONNXRUNTIME_ROOT}/core/providers/rocm/rocm_stream_handle.cc"
|
||||
)
|
||||
source_group(TREE ${ONNXRUNTIME_ROOT}/core FILES ${onnxruntime_providers_migraphx_cc_srcs})
|
||||
onnxruntime_add_shared_library_module(onnxruntime_providers_migraphx ${onnxruntime_providers_migraphx_cc_srcs})
|
||||
|
|
@ -48,16 +46,18 @@
|
|||
set_target_properties(onnxruntime_providers_migraphx PROPERTIES LINKER_LANGUAGE CXX)
|
||||
set_target_properties(onnxruntime_providers_migraphx PROPERTIES FOLDER "ONNXRuntime")
|
||||
target_compile_definitions(onnxruntime_providers_migraphx PRIVATE ONNXIFI_BUILD_LIBRARY=1)
|
||||
if(MSVC)
|
||||
set_property(TARGET onnxruntime_providers_migraphx APPEND_STRING PROPERTY LINK_FLAGS /DEF:${ONNXRUNTIME_ROOT}/core/providers/migraphx/symbols.def)
|
||||
target_link_libraries(onnxruntime_providers_migraphx PRIVATE ws2_32)
|
||||
target_compile_options(onnxruntime_providers_migraphx PRIVATE -Wno-error=sign-compare)
|
||||
set_property(TARGET onnxruntime_providers_migraphx APPEND_STRING PROPERTY COMPILE_FLAGS "-Wno-deprecated-declarations")
|
||||
set_property(TARGET onnxruntime_providers_migraphx APPEND_STRING PROPERTY LINK_FLAGS "-Xlinker --version-script=${ONNXRUNTIME_ROOT}/core/providers/migraphx/version_script.lds -Xlinker --gc-sections")
|
||||
target_link_libraries(onnxruntime_providers_migraphx PRIVATE nsync::nsync_cpp)
|
||||
|
||||
include(CheckLibraryExists)
|
||||
check_library_exists(migraphx::c "migraphx_program_run_async" "/opt/rocm/migraphx/lib" HAS_STREAM_SYNC)
|
||||
if(HAS_STREAM_SYNC)
|
||||
target_compile_definitions(onnxruntime_providers_migraphx PRIVATE -DMIGRAPHX_STREAM_SYNC)
|
||||
message(STATUS "MIGRAPHX GPU STREAM SYNC is ENABLED")
|
||||
else()
|
||||
target_compile_options(onnxruntime_providers_migraphx PRIVATE -Wno-error=sign-compare)
|
||||
set_property(TARGET onnxruntime_providers_migraphx APPEND_STRING PROPERTY COMPILE_FLAGS "-Wno-deprecated-declarations")
|
||||
endif()
|
||||
if(UNIX)
|
||||
set_property(TARGET onnxruntime_providers_migraphx APPEND_STRING PROPERTY LINK_FLAGS "-Xlinker --version-script=${ONNXRUNTIME_ROOT}/core/providers/migraphx/version_script.lds -Xlinker --gc-sections")
|
||||
target_link_libraries(onnxruntime_providers_migraphx PRIVATE nsync::nsync_cpp stdc++fs)
|
||||
message(STATUS "MIGRAPHX GPU STREAM SYNC is DISABLED")
|
||||
endif()
|
||||
|
||||
if (onnxruntime_ENABLE_TRAINING_OPS)
|
||||
|
|
@ -68,16 +68,8 @@
|
|||
endif()
|
||||
endif()
|
||||
|
||||
if(CMAKE_SYSTEM_NAME STREQUAL "Windows")
|
||||
install(TARGETS onnxruntime_providers_migraphx
|
||||
ARCHIVE DESTINATION ${CMAKE_INSTALL_LIBDIR}
|
||||
LIBRARY DESTINATION ${CMAKE_INSTALL_BINDIR}
|
||||
RUNTIME DESTINATION ${CMAKE_INSTALL_BINDIR}
|
||||
)
|
||||
else()
|
||||
install(TARGETS onnxruntime_providers_migraphx
|
||||
ARCHIVE DESTINATION ${CMAKE_INSTALL_LIBDIR}
|
||||
LIBRARY DESTINATION ${CMAKE_INSTALL_LIBDIR}
|
||||
RUNTIME DESTINATION ${CMAKE_INSTALL_BINDIR}
|
||||
)
|
||||
endif()
|
||||
install(TARGETS onnxruntime_providers_migraphx
|
||||
ARCHIVE DESTINATION ${CMAKE_INSTALL_LIBDIR}
|
||||
LIBRARY DESTINATION ${CMAKE_INSTALL_LIBDIR}
|
||||
RUNTIME DESTINATION ${CMAKE_INSTALL_BINDIR}
|
||||
)
|
||||
|
|
|
|||
|
|
@ -60,7 +60,17 @@ common::Status GPUDataTransfer::CopyTensorAsync(const Tensor& src, Tensor& dst,
|
|||
HIP_CALL_THROW(hipMemcpy(dst_data, src_data, bytes, hipMemcpyHostToDevice));
|
||||
}
|
||||
} else if (src_device.Type() == OrtDevice::GPU) {
|
||||
#ifndef MIGRAPHX_STREAM_SYNC
|
||||
if (dst_device.Type() == OrtDevice::CPU && dst_device.MemType() == OrtDevice::MemType::HIP_PINNED) {
|
||||
// copying from GPU to pinned memory, this is non-blocking
|
||||
HIP_CALL_THROW(hipMemcpyAsync(dst_data, src_data, bytes, hipMemcpyDeviceToHost, static_cast<hipStream_t>(stream.GetHandle())));
|
||||
} else {
|
||||
// copying from GPU to CPU memory, this is blocking
|
||||
HIP_CALL_THROW(hipMemcpy(dst_data, src_data, bytes, hipMemcpyDeviceToHost));
|
||||
}
|
||||
#else
|
||||
HIP_CALL_THROW(hipMemcpyAsync(dst_data, src_data, bytes, hipMemcpyDeviceToHost, static_cast<hipStream_t>(stream.GetHandle())));
|
||||
#endif
|
||||
} else {
|
||||
// copying between cpu memory
|
||||
memcpy(dst_data, src_data, bytes);
|
||||
|
|
|
|||
|
|
@ -3,7 +3,7 @@
|
|||
|
||||
#include "core/providers/shared_library/provider_api.h"
|
||||
#include "migraphx_call.h"
|
||||
#include "migraphx_allocator.h"
|
||||
#include "hip_allocator.h"
|
||||
#include "core/common/status.h"
|
||||
#include "core/framework/float16.h"
|
||||
#include "core/common/status.h"
|
||||
|
|
@ -11,7 +11,7 @@
|
|||
|
||||
namespace onnxruntime {
|
||||
|
||||
void MIGraphXAllocator::CheckDevice() const {
|
||||
void HIPAllocator::CheckDevice() const {
|
||||
#ifndef NDEBUG
|
||||
// check device to match at debug build
|
||||
// if it's expected to change, call hipSetDevice instead of the check
|
||||
|
|
@ -23,7 +23,7 @@ void MIGraphXAllocator::CheckDevice() const {
|
|||
#endif
|
||||
}
|
||||
|
||||
void* MIGraphXAllocator::Alloc(size_t size) {
|
||||
void* HIPAllocator::Alloc(size_t size) {
|
||||
CheckDevice();
|
||||
void* p = nullptr;
|
||||
if (size > 0) {
|
||||
|
|
@ -32,12 +32,12 @@ void* MIGraphXAllocator::Alloc(size_t size) {
|
|||
return p;
|
||||
}
|
||||
|
||||
void MIGraphXAllocator::Free(void* p) {
|
||||
void HIPAllocator::Free(void* p) {
|
||||
CheckDevice();
|
||||
(void)hipFree(p); // do not throw error since it's OK for hipFree to fail during shutdown
|
||||
}
|
||||
|
||||
void* MIGraphXExternalAllocator::Alloc(size_t size) {
|
||||
void* HIPExternalAllocator::Alloc(size_t size) {
|
||||
void* p = nullptr;
|
||||
if (size > 0) {
|
||||
p = alloc_(size);
|
||||
|
|
@ -49,7 +49,7 @@ void* MIGraphXExternalAllocator::Alloc(size_t size) {
|
|||
return p;
|
||||
}
|
||||
|
||||
void MIGraphXExternalAllocator::Free(void* p) {
|
||||
void HIPExternalAllocator::Free(void* p) {
|
||||
free_(p);
|
||||
std::lock_guard<OrtMutex> lock(lock_);
|
||||
auto it = reserved_.find(p);
|
||||
|
|
@ -59,7 +59,7 @@ void MIGraphXExternalAllocator::Free(void* p) {
|
|||
}
|
||||
}
|
||||
|
||||
void* MIGraphXExternalAllocator::Reserve(size_t size) {
|
||||
void* HIPExternalAllocator::Reserve(size_t size) {
|
||||
void* p = Alloc(size);
|
||||
if (!p) return nullptr;
|
||||
std::lock_guard<OrtMutex> lock(lock_);
|
||||
|
|
@ -9,12 +9,12 @@
|
|||
|
||||
namespace onnxruntime {
|
||||
|
||||
class MIGraphXAllocator : public IAllocator {
|
||||
class HIPAllocator : public IAllocator {
|
||||
public:
|
||||
MIGraphXAllocator(int device_id, const char* name)
|
||||
HIPAllocator(int device_id, const char* name)
|
||||
: IAllocator(
|
||||
OrtMemoryInfo(name, OrtAllocatorType::OrtDeviceAllocator,
|
||||
OrtDevice(OrtDevice::GPU, OrtDevice::MemType::DEFAULT, static_cast<OrtDevice::DeviceId>(device_id)),
|
||||
OrtDevice(OrtDevice::GPU, OrtDevice::MemType::DEFAULT, device_id),
|
||||
device_id, OrtMemTypeDefault)) {}
|
||||
|
||||
virtual void* Alloc(size_t size) override;
|
||||
|
|
@ -24,14 +24,14 @@ class MIGraphXAllocator : public IAllocator {
|
|||
void CheckDevice() const;
|
||||
};
|
||||
|
||||
class MIGraphXExternalAllocator : public MIGraphXAllocator {
|
||||
class HIPExternalAllocator : public HIPAllocator {
|
||||
typedef void* (*ExternalAlloc)(size_t size);
|
||||
typedef void (*ExternalFree)(void* p);
|
||||
typedef void (*ExternalEmptyCache)();
|
||||
|
||||
public:
|
||||
MIGraphXExternalAllocator(OrtDevice::DeviceId device_id, const char* name, void* alloc, void* free, void* empty_cache)
|
||||
: MIGraphXAllocator(device_id, name) {
|
||||
HIPExternalAllocator(OrtDevice::DeviceId device_id, const char* name, void* alloc, void* free, void* empty_cache)
|
||||
: HIPAllocator(device_id, name) {
|
||||
alloc_ = reinterpret_cast<ExternalAlloc>(alloc);
|
||||
free_ = reinterpret_cast<ExternalFree>(free);
|
||||
empty_cache_ = reinterpret_cast<ExternalEmptyCache>(empty_cache);
|
||||
|
|
@ -55,7 +55,7 @@ class HIPPinnedAllocator : public IAllocator {
|
|||
HIPPinnedAllocator(int device_id, const char* name)
|
||||
: IAllocator(
|
||||
OrtMemoryInfo(name, OrtAllocatorType::OrtDeviceAllocator,
|
||||
OrtDevice(OrtDevice::CPU, OrtDevice::MemType::HIP_PINNED, static_cast<OrtDevice::DeviceId>(device_id)),
|
||||
OrtDevice(OrtDevice::CPU, OrtDevice::MemType::HIP_PINNED, device_id),
|
||||
device_id, OrtMemTypeCPUOutput)) {}
|
||||
|
||||
virtual void* Alloc(size_t size) override;
|
||||
|
|
@ -1,13 +1,10 @@
|
|||
// Copyright (c) Microsoft Corporation. All rights reserved.
|
||||
// Licensed under the MIT License.
|
||||
|
||||
#ifdef _WIN32
|
||||
#include <winsock.h>
|
||||
#else
|
||||
#include <unistd.h>
|
||||
#endif
|
||||
|
||||
#include <string>
|
||||
#include <string.h>
|
||||
#include <miopen/miopen.h>
|
||||
#include <rocblas/rocblas.h>
|
||||
#include "core/common/common.h"
|
||||
#include "core/common/status.h"
|
||||
#include "core/providers/shared_library/provider_api.h"
|
||||
|
|
@ -37,20 +34,16 @@ std::conditional_t<THRW, void, Status> RocmCall(
|
|||
ERRTYPE retCode, const char* exprString, const char* libName, ERRTYPE successCode, const char* msg, const char* file, const int line) {
|
||||
if (retCode != successCode) {
|
||||
try {
|
||||
#ifdef _WIN32
|
||||
// According to the POSIX spec, 255 is the safe minimum value.
|
||||
static constexpr int HOST_NAME_MAX = 255;
|
||||
#endif
|
||||
std::string hostname(HOST_NAME_MAX, 0);
|
||||
if (gethostname(hostname.data(), HOST_NAME_MAX) != 0)
|
||||
hostname = "?";
|
||||
char hostname[HOST_NAME_MAX];
|
||||
if (gethostname(hostname, HOST_NAME_MAX) != 0)
|
||||
strcpy(hostname, "?");
|
||||
int currentHipDevice;
|
||||
(void)hipGetDevice(¤tHipDevice);
|
||||
(void)hipGetLastError(); // clear last HIP error
|
||||
static char str[1024];
|
||||
snprintf(str, 1024, "%s failure %d: %s ; GPU=%d ; hostname=%s ; file=%s ; line=%d ; expr=%s; %s",
|
||||
libName, (int)retCode, RocmErrString(retCode), currentHipDevice,
|
||||
hostname.c_str(),
|
||||
hostname,
|
||||
file, line, exprString, msg);
|
||||
if constexpr (THRW) {
|
||||
// throw an exception with the error info
|
||||
|
|
@ -75,5 +68,9 @@ std::conditional_t<THRW, void, Status> RocmCall(
|
|||
|
||||
template Status RocmCall<hipError_t, false>(hipError_t retCode, const char* exprString, const char* libName, hipError_t successCode, const char* msg, const char* file, const int line);
|
||||
template void RocmCall<hipError_t, true>(hipError_t retCode, const char* exprString, const char* libName, hipError_t successCode, const char* msg, const char* file, const int line);
|
||||
template Status RocmCall<rocblas_status, false>(rocblas_status retCode, const char* exprString, const char* libName, rocblas_status successCode, const char* msg, const char* file, const int line);
|
||||
template void RocmCall<rocblas_status, true>(rocblas_status retCode, const char* exprString, const char* libName, rocblas_status successCode, const char* msg, const char* file, const int line);
|
||||
template Status RocmCall<miopenStatus_t, false>(miopenStatus_t retCode, const char* exprString, const char* libName, miopenStatus_t successCode, const char* msg, const char* file, const int line);
|
||||
template void RocmCall<miopenStatus_t, true>(miopenStatus_t retCode, const char* exprString, const char* libName, miopenStatus_t successCode, const char* msg, const char* file, const int line);
|
||||
|
||||
} // namespace onnxruntime
|
||||
|
|
|
|||
|
|
@ -4,6 +4,8 @@
|
|||
#pragma once
|
||||
#include "migraphx_inc.h"
|
||||
|
||||
#pragma once
|
||||
|
||||
namespace onnxruntime {
|
||||
|
||||
// -----------------------------------------------------------------------
|
||||
|
|
|
|||
|
|
@ -13,11 +13,12 @@
|
|||
#include "core/common/logging/severity.h"
|
||||
#include "migraphx_execution_provider.h"
|
||||
#include "migraphx_execution_provider_utils.h"
|
||||
#include "migraphx_allocator.h"
|
||||
#include "hip_allocator.h"
|
||||
#include "gpu_data_transfer.h"
|
||||
#include "migraphx_inc.h"
|
||||
|
||||
#include "migraphx_stream_handle.h"
|
||||
// TODO: find a better way to share this
|
||||
#include "core/providers/rocm/rocm_stream_handle.h"
|
||||
|
||||
#if defined(_MSC_VER)
|
||||
#pragma warning(disable : 4244 4245)
|
||||
|
|
@ -101,10 +102,10 @@ std::shared_ptr<KernelRegistry> MIGraphXExecutionProvider::GetKernelRegistry() c
|
|||
}
|
||||
|
||||
MIGraphXExecutionProvider::MIGraphXExecutionProvider(const MIGraphXExecutionProviderInfo& info)
|
||||
: IExecutionProvider{onnxruntime::kMIGraphXExecutionProvider, OrtDevice(OrtDevice::GPU, OrtDevice::MemType::DEFAULT, info.device_id)}, info_(info) {
|
||||
: IExecutionProvider{onnxruntime::kMIGraphXExecutionProvider, OrtDevice(OrtDevice::GPU, OrtDevice::MemType::DEFAULT, info.device_id)}, device_id_(info.device_id) {
|
||||
InitProviderOrtApi();
|
||||
// Set GPU device to be used
|
||||
HIP_CALL_THROW(hipSetDevice(info_.device_id));
|
||||
HIP_CALL_THROW(hipSetDevice(device_id_));
|
||||
t_ = migraphx::target(info.target_device.c_str());
|
||||
|
||||
// whether fp16 is enable
|
||||
|
|
@ -180,10 +181,16 @@ MIGraphXExecutionProvider::MIGraphXExecutionProvider(const MIGraphXExecutionProv
|
|||
dump_model_ops_ = (std::stoi(dump_model_ops_env) == 0 ? false : true);
|
||||
}
|
||||
|
||||
ROCBLAS_CALL_THROW(rocblas_create_handle(&external_rocblas_handle_));
|
||||
ROCBLAS_CALL_THROW(rocblas_set_stream(external_rocblas_handle_, stream_));
|
||||
|
||||
MIOPEN_CALL_THROW(miopenCreate(&external_miopen_handle_));
|
||||
MIOPEN_CALL_THROW(miopenSetStream(external_miopen_handle_, stream_));
|
||||
|
||||
metadef_id_generator_ = ModelMetadefIdGenerator::Create();
|
||||
|
||||
LOGS_DEFAULT(VERBOSE) << "[MIGraphX EP] MIGraphX provider options: "
|
||||
<< "device_id: " << info_.device_id
|
||||
<< "device_id: " << device_id_
|
||||
<< ", migraphx_fp16_enable: " << fp16_enable_
|
||||
<< ", migraphx_int8_enable: " << int8_enable_
|
||||
<< ", migraphx_int8_enable: " << int8_enable_
|
||||
|
|
@ -198,14 +205,17 @@ MIGraphXExecutionProvider::MIGraphXExecutionProvider(const MIGraphXExecutionProv
|
|||
}
|
||||
|
||||
MIGraphXExecutionProvider::~MIGraphXExecutionProvider() {
|
||||
ORT_IGNORE_RETURN_VALUE(ROCBLAS_CALL(rocblas_destroy_handle(external_rocblas_handle_)));
|
||||
ORT_IGNORE_RETURN_VALUE(MIOPEN_CALL(miopenDestroy(external_miopen_handle_)));
|
||||
}
|
||||
|
||||
std::vector<AllocatorPtr> MIGraphXExecutionProvider::CreatePreferredAllocators() {
|
||||
AllocatorCreationInfo default_memory_info(
|
||||
[](OrtDevice::DeviceId device_id) { return CreateMIGraphXAllocator(device_id, onnxruntime::CUDA); }, info_.device_id);
|
||||
[](OrtDevice::DeviceId device_id) { return CreateROCMAllocator(device_id, onnxruntime::CUDA); }, device_id_);
|
||||
AllocatorCreationInfo pinned_allocator_info(
|
||||
[](OrtDevice::DeviceId device_id) {
|
||||
return CreateMIGraphXPinnedAllocator(device_id, onnxruntime::CUDA_PINNED);
|
||||
ORT_UNUSED_PARAMETER(device_id);
|
||||
return CreateROCMPinnedAllocator(onnxruntime::CUDA_PINNED);
|
||||
},
|
||||
0);
|
||||
return std::vector<AllocatorPtr>{CreateAllocator(default_memory_info), CreateAllocator(pinned_allocator_info)};
|
||||
|
|
@ -244,40 +254,40 @@ static bool getMIGraphXType(ONNXTensorElementDataType type,
|
|||
migraphx_shape_datatype_t& mgx_type) {
|
||||
mgx_type = migraphx_shape_float_type;
|
||||
switch (type) {
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT16:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_FLOAT16:
|
||||
mgx_type = migraphx_shape_half_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_FLOAT:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_FLOAT:
|
||||
mgx_type = migraphx_shape_float_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_DOUBLE:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_DOUBLE:
|
||||
mgx_type = migraphx_shape_double_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_INT8:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_INT8:
|
||||
mgx_type = migraphx_shape_int8_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_INT16:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_INT16:
|
||||
mgx_type = migraphx_shape_int16_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_INT32:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_INT32:
|
||||
mgx_type = migraphx_shape_int32_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_INT64:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_INT64:
|
||||
mgx_type = migraphx_shape_int64_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_UINT8:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_UINT8:
|
||||
mgx_type = migraphx_shape_uint8_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_UINT16:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_UINT16:
|
||||
mgx_type = migraphx_shape_uint16_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_UINT32:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_UINT32:
|
||||
mgx_type = migraphx_shape_uint32_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_UINT64:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_UINT64:
|
||||
mgx_type = migraphx_shape_uint64_type;
|
||||
break;
|
||||
case ONNX_TENSOR_ELEMENT_DATA_TYPE_BOOL:
|
||||
case ONNX_NAMESPACE::TensorProto_DataType::TensorProto_DataType_BOOL:
|
||||
mgx_type = migraphx_shape_bool_type;
|
||||
break;
|
||||
default:
|
||||
|
|
@ -293,7 +303,7 @@ std::vector<int> toVector(const ONNX_NAMESPACE::int64s& nums) {
|
|||
std::vector<int> result;
|
||||
int num = nums.size();
|
||||
for (int i = 0; i < num; ++i) {
|
||||
result.push_back(static_cast<int>(nums[i]));
|
||||
result.push_back(nums[i]);
|
||||
}
|
||||
|
||||
return result;
|
||||
|
|
@ -491,9 +501,16 @@ static bool IsUnsupportedOpMode(const onnxruntime::GraphViewer& graph_viewer, co
|
|||
if (arg_s != nullptr) {
|
||||
const auto& tensor_dims = arg_s->dim();
|
||||
std::vector<std::size_t> dims;
|
||||
for (auto&& dim : tensor_dims) {
|
||||
dims.emplace_back(dim.has_dim_value() ? dim.dim_value() : 0);
|
||||
}
|
||||
std::transform(tensor_dims.begin(),
|
||||
tensor_dims.end(),
|
||||
std::back_inserter(dims),
|
||||
[&](auto&& d) -> std::size_t {
|
||||
if (d.has_dim_value()) {
|
||||
return d.dim_value();
|
||||
} else {
|
||||
return 0;
|
||||
}
|
||||
});
|
||||
if (dims == std::vector<std::size_t>{0}) {
|
||||
return true;
|
||||
}
|
||||
|
|
@ -529,8 +546,8 @@ static bool IsUnsupportedOpMode(const onnxruntime::GraphViewer& graph_viewer, co
|
|||
}
|
||||
|
||||
void SubgraphPostProcessing(const onnxruntime::GraphViewer& graph_viewer, std::vector<std::vector<NodeIndex>>& clusters,
|
||||
[[maybe_unused]] const logging::Logger& logger) {
|
||||
// Then check whether a subgraph should fall back to CPU
|
||||
const logging::Logger& logger) {
|
||||
// Then check whether a subgraph should fallback to CPU
|
||||
// 1. Check whether a subgraph contains a RNN operator
|
||||
std::unordered_set<std::string> rnn_names = {"RNN", "GRU", "LSTM"};
|
||||
std::unordered_set<std::string> op_names = {"AveragePool", "Conv", "Gemm", "LRN", "MatMul", "MaxPool"};
|
||||
|
|
@ -574,10 +591,17 @@ void SubgraphPostProcessing(const onnxruntime::GraphViewer& graph_viewer, std::v
|
|||
if (arg_s == nullptr) return false;
|
||||
const auto& tensor_dims = arg_s->dim();
|
||||
std::vector<std::size_t> dims;
|
||||
for (auto&& dim : tensor_dims) {
|
||||
dims.emplace_back(dim.has_dim_value() ? dim.dim_value() : 1);
|
||||
}
|
||||
return (std::accumulate(dims.begin(), dims.end(), 1ULL, std::multiplies<std::size_t>{}) > 300);
|
||||
std::transform(tensor_dims.begin(),
|
||||
tensor_dims.end(),
|
||||
std::back_inserter(dims),
|
||||
[&](auto&& d) -> std::size_t {
|
||||
if (d.has_dim_value()) {
|
||||
return d.dim_value();
|
||||
} else {
|
||||
return 1;
|
||||
}
|
||||
});
|
||||
return (std::accumulate(dims.begin(), dims.end(), 1, std::multiplies<std::size_t>{}) > 300);
|
||||
})) {
|
||||
return false;
|
||||
}
|
||||
|
|
@ -599,7 +623,7 @@ void SubgraphPostProcessing(const onnxruntime::GraphViewer& graph_viewer, std::v
|
|||
static bool IsNodeSupported(const std::set<std::string>& op_set,
|
||||
const onnxruntime::GraphViewer& graph_viewer,
|
||||
const NodeIndex node_idx,
|
||||
[[maybe_unused]] const logging::Logger& logger) {
|
||||
const logging::Logger& logger) {
|
||||
const auto& node = graph_viewer.GetNode(node_idx);
|
||||
const auto& optype = node->OpType();
|
||||
const auto& domain = node->Domain();
|
||||
|
|
@ -1418,10 +1442,14 @@ Status MIGraphXExecutionProvider::Compile(const std::vector<FusedNodeAndGraph>&
|
|||
// lock to avoid race condition
|
||||
std::lock_guard<OrtMutex> lock(*(mgx_state->mgx_mu_ptr));
|
||||
|
||||
#ifdef MIGRAPHX_STREAM_SYNC
|
||||
void* rocm_stream;
|
||||
Ort::ThrowOnError(api->KernelContext_GetGPUComputeStream(context, &rocm_stream));
|
||||
auto prog_outputs = prog.run_async(m, static_cast<hipStream_t>(rocm_stream));
|
||||
|
||||
#else
|
||||
auto prog_outputs = prog.eval(m);
|
||||
HIP_CALL_THROW(hipDeviceSynchronize());
|
||||
#endif
|
||||
// In case of input parameters are reused as output parameter call hipMemcpy
|
||||
auto output_num = prog_outputs.size();
|
||||
if (prog_output_indices.size() < output_num) {
|
||||
|
|
@ -1450,7 +1478,8 @@ Status MIGraphXExecutionProvider::Compile(const std::vector<FusedNodeAndGraph>&
|
|||
void MIGraphXExecutionProvider::RegisterStreamHandlers(IStreamCommandHandleRegistry& stream_handle_registry,
|
||||
AllocatorMap& allocators) const {
|
||||
auto allocator = allocators[GetOrtDeviceByMemType(OrtMemTypeCPU)];
|
||||
RegisterMIGraphXStreamHandles(stream_handle_registry, OrtDevice::GPU, allocator, true, stream_, false /*TODO:external_stream_*/);
|
||||
RegisterRocmStreamHandles(stream_handle_registry, OrtDevice::GPU, allocator, true, stream_,
|
||||
false /*TODO:external_stream_*/, external_miopen_handle_, external_rocblas_handle_);
|
||||
}
|
||||
|
||||
OrtDevice MIGraphXExecutionProvider::GetOrtDeviceByMemType(OrtMemType mem_type) const {
|
||||
|
|
@ -1458,6 +1487,7 @@ OrtDevice MIGraphXExecutionProvider::GetOrtDeviceByMemType(OrtMemType mem_type)
|
|||
if (mem_type == OrtMemTypeCPUOutput) return OrtDevice(OrtDevice::CPU, OrtDevice::MemType::HIP_PINNED, 0 /*CPU device id always be 0*/);
|
||||
return default_device_;
|
||||
}
|
||||
#ifdef MIGRAPHX_STREAM_SYNC
|
||||
|
||||
Status MIGraphXExecutionProvider::Sync() const {
|
||||
HIP_CALL_THROW(hipStreamSynchronize(static_cast<hipStream_t>(nullptr)));
|
||||
|
|
@ -1482,4 +1512,5 @@ Status MIGraphXExecutionProvider::OnRunEnd(bool /*sync_stream*/, const onnxrunti
|
|||
return Status::OK();
|
||||
}
|
||||
|
||||
#endif
|
||||
} // namespace onnxruntime
|
||||
|
|
|
|||
|
|
@ -3,6 +3,9 @@
|
|||
|
||||
#pragma once
|
||||
|
||||
#include <miopen/miopen.h>
|
||||
#include <rocblas/rocblas.h>
|
||||
|
||||
#include "core/framework/arena_extend_strategy.h"
|
||||
#include "core/framework/execution_provider.h"
|
||||
#include "core/platform/ort_mutex.h"
|
||||
|
|
@ -11,6 +14,8 @@
|
|||
|
||||
#include <map>
|
||||
#include <unordered_map>
|
||||
// TODO: find a better way to share this
|
||||
// #include "core/providers/cuda/rocm_stream_handle.h"
|
||||
|
||||
namespace onnxruntime {
|
||||
|
||||
|
|
@ -57,11 +62,13 @@ class MIGraphXExecutionProvider : public IExecutionProvider {
|
|||
explicit MIGraphXExecutionProvider(const MIGraphXExecutionProviderInfo& info);
|
||||
~MIGraphXExecutionProvider();
|
||||
|
||||
#ifdef MIGRAPHX_STREAM_SYNC
|
||||
Status Sync() const override;
|
||||
|
||||
Status OnRunStart(const onnxruntime::RunOptions& run_options) override;
|
||||
|
||||
Status OnRunEnd(bool sync_stream, const onnxruntime::RunOptions& run_options) override;
|
||||
#endif
|
||||
|
||||
std::vector<std::unique_ptr<ComputeCapability>>
|
||||
GetCapability(const onnxruntime::GraphViewer& graph_viewer,
|
||||
|
|
@ -78,13 +85,7 @@ class MIGraphXExecutionProvider : public IExecutionProvider {
|
|||
OrtDevice GetOrtDeviceByMemType(OrtMemType mem_type) const override;
|
||||
std::vector<AllocatorPtr> CreatePreferredAllocators() override;
|
||||
|
||||
int GetDeviceId() const override { return info_.device_id; }
|
||||
ProviderOptions GetProviderOptions() const override {
|
||||
return MIGraphXExecutionProviderInfo::ToProviderOptions(info_);
|
||||
}
|
||||
|
||||
private:
|
||||
MIGraphXExecutionProviderInfo info_;
|
||||
bool fp16_enable_ = false;
|
||||
bool int8_enable_ = false;
|
||||
std::string int8_calibration_cache_name_;
|
||||
|
|
@ -97,6 +98,7 @@ class MIGraphXExecutionProvider : public IExecutionProvider {
|
|||
bool load_compiled_model_ = false;
|
||||
std::string load_compiled_path_;
|
||||
bool dump_model_ops_ = false;
|
||||
int device_id_;
|
||||
migraphx::target t_;
|
||||
OrtMutex mgx_mu_;
|
||||
hipStream_t stream_ = nullptr;
|
||||
|
|
@ -107,6 +109,8 @@ class MIGraphXExecutionProvider : public IExecutionProvider {
|
|||
std::unordered_map<std::string, bool> map_no_input_shape_;
|
||||
|
||||
AllocatorPtr allocator_;
|
||||
miopenHandle_t external_miopen_handle_ = nullptr;
|
||||
rocblas_handle external_rocblas_handle_ = nullptr;
|
||||
std::unique_ptr<ModelMetadefIdGenerator> metadef_id_generator_;
|
||||
};
|
||||
|
||||
|
|
|
|||
|
|
@ -14,7 +14,7 @@ namespace onnxruntime {
|
|||
// Information needed to construct trt execution providers.
|
||||
struct MIGraphXExecutionProviderInfo {
|
||||
std::string target_device;
|
||||
OrtDevice::DeviceId device_id{0};
|
||||
int device_id{0};
|
||||
bool fp16_enable{false};
|
||||
bool int8_enable{false};
|
||||
std::string int8_calibration_table_name{""};
|
||||
|
|
|
|||
|
|
@ -28,7 +28,7 @@ bool IsGraphInput(const GraphViewer& graph, const std::string& name) {
|
|||
return (std::find(input_names.begin(), input_names.end(), name) != input_names.end());
|
||||
}
|
||||
|
||||
bool IsGraphInitializer(const GraphViewer& graph, const std::string& name, [[maybe_unused]] bool check_outer_scope = true) {
|
||||
bool IsGraphInitializer(const GraphViewer& graph, const std::string& name, bool check_outer_scope = true) {
|
||||
const ONNX_NAMESPACE::TensorProto* initializer = nullptr;
|
||||
return graph.GetInitializedTensor(name, initializer);
|
||||
}
|
||||
|
|
|
|||
|
|
@ -4,5 +4,5 @@
|
|||
#pragma once
|
||||
|
||||
#include <hip/hip_runtime.h>
|
||||
#include <iso646.h>
|
||||
#include <migraphx/migraphx.h>
|
||||
#include <migraphx/migraphx.hpp>
|
||||
|
|
|
|||
|
|
@ -6,7 +6,7 @@
|
|||
#include "core/providers/migraphx/migraphx_provider_factory.h"
|
||||
#include "migraphx_execution_provider.h"
|
||||
#include "migraphx_provider_factory_creator.h"
|
||||
#include "migraphx_allocator.h"
|
||||
#include "hip_allocator.h"
|
||||
#include "gpu_data_transfer.h"
|
||||
#include "core/framework/provider_options.h"
|
||||
|
||||
|
|
@ -33,23 +33,10 @@ std::unique_ptr<IExecutionProvider> MIGraphXProviderFactory::CreateProvider() {
|
|||
return std::make_unique<MIGraphXExecutionProvider>(info_);
|
||||
}
|
||||
|
||||
struct ProviderInfo_MIGraphX_Impl final : ProviderInfo_MIGraphX {
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXAllocator(int16_t device_id, const char* name) override {
|
||||
return std::make_unique<MIGraphXAllocator>(device_id, name);
|
||||
}
|
||||
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXPinnedAllocator(int16_t device_id, const char* name) override {
|
||||
return std::make_unique<HIPPinnedAllocator>(device_id, name);
|
||||
}
|
||||
|
||||
} g_info;
|
||||
|
||||
struct MIGraphX_Provider : Provider {
|
||||
void* GetInfo() override { return &g_info; }
|
||||
|
||||
std::shared_ptr<IExecutionProviderFactory> CreateExecutionProviderFactory(int device_id) override {
|
||||
MIGraphXExecutionProviderInfo info;
|
||||
info.device_id = static_cast<OrtDevice::DeviceId>(device_id);
|
||||
info.device_id = device_id;
|
||||
info.target_device = "gpu";
|
||||
return std::make_shared<MIGraphXProviderFactory>(info);
|
||||
}
|
||||
|
|
@ -57,7 +44,7 @@ struct MIGraphX_Provider : Provider {
|
|||
std::shared_ptr<IExecutionProviderFactory> CreateExecutionProviderFactory(const void* provider_options) override {
|
||||
auto& options = *reinterpret_cast<const OrtMIGraphXProviderOptions*>(provider_options);
|
||||
MIGraphXExecutionProviderInfo info;
|
||||
info.device_id = static_cast<OrtDevice::DeviceId>(options.device_id);
|
||||
info.device_id = options.device_id;
|
||||
info.target_device = "gpu";
|
||||
info.fp16_enable = options.migraphx_fp16_enable;
|
||||
info.int8_enable = options.migraphx_int8_enable;
|
||||
|
|
|
|||
|
|
@ -10,13 +10,4 @@ struct IExecutionProviderFactory;
|
|||
struct MIGraphXExecutionProviderInfo;
|
||||
enum class ArenaExtendStrategy : int32_t;
|
||||
struct MIGraphXExecutionProviderExternalAllocatorInfo;
|
||||
|
||||
struct ProviderInfo_MIGraphX {
|
||||
virtual std::unique_ptr<onnxruntime::IAllocator> CreateMIGraphXAllocator(int16_t device_id, const char* name) = 0;
|
||||
virtual std::unique_ptr<onnxruntime::IAllocator> CreateMIGraphXPinnedAllocator(int16_t device_id, const char* name) = 0;
|
||||
|
||||
protected:
|
||||
~ProviderInfo_MIGraphX() = default; // Can only be destroyed through a subclass instance
|
||||
};
|
||||
|
||||
} // namespace onnxruntime
|
||||
|
|
|
|||
|
|
@ -1,171 +0,0 @@
|
|||
// Copyright (c) Microsoft Corporation. All rights reserved.
|
||||
// Licensed under the MIT License.
|
||||
|
||||
#include <core/providers/rocm/rocm_resource.h>
|
||||
#include "migraphx_stream_handle.h"
|
||||
|
||||
namespace onnxruntime {
|
||||
|
||||
struct MIGraphXNotification : public synchronize::Notification {
|
||||
MIGraphXNotification(Stream& s) : Notification(s) {
|
||||
HIP_CALL_THROW(hipEventCreateWithFlags(&event_, hipEventDisableTiming));
|
||||
}
|
||||
|
||||
~MIGraphXNotification() {
|
||||
if (event_)
|
||||
HIP_CALL_THROW(hipEventDestroy(event_));
|
||||
}
|
||||
|
||||
void Activate() override {
|
||||
// record event with hipEventBlockingSync so we can support sync on host without busy wait.
|
||||
HIP_CALL_THROW(hipEventRecord(event_, static_cast<hipStream_t>(stream_.GetHandle())));
|
||||
}
|
||||
|
||||
void wait_on_device(Stream& device_stream) {
|
||||
ORT_ENFORCE(device_stream.GetDevice().Type() == OrtDevice::GPU, "Unexpected device:", device_stream.GetDevice().ToString());
|
||||
// launch a wait command to the migraphx stream
|
||||
HIP_CALL_THROW(hipStreamWaitEvent(static_cast<hipStream_t>(device_stream.GetHandle()), event_, 0));
|
||||
};
|
||||
|
||||
void wait_on_host() {
|
||||
// CUDA_CALL_THROW(cudaStreamSynchronize(stream_));
|
||||
HIP_CALL_THROW(hipEventSynchronize(event_));
|
||||
}
|
||||
|
||||
hipEvent_t event_;
|
||||
};
|
||||
|
||||
MIGraphXStream::MIGraphXStream(hipStream_t stream,
|
||||
const OrtDevice& device,
|
||||
AllocatorPtr cpu_allocator,
|
||||
bool release_cpu_buffer_on_migraphx_stream)
|
||||
: Stream(stream, device),
|
||||
cpu_allocator_(cpu_allocator),
|
||||
release_cpu_buffer_on_migraphx_stream_(release_cpu_buffer_on_migraphx_stream) {
|
||||
}
|
||||
|
||||
MIGraphXStream::~MIGraphXStream() {
|
||||
ORT_IGNORE_RETURN_VALUE(CleanUpOnRunEnd());
|
||||
if (own_stream_) {
|
||||
auto* handle = GetHandle();
|
||||
if (handle)
|
||||
HIP_CALL_THROW(hipStreamDestroy(static_cast<hipStream_t>(handle)));
|
||||
}
|
||||
}
|
||||
|
||||
std::unique_ptr<synchronize::Notification> MIGraphXStream::CreateNotification(size_t /*num_consumers*/) {
|
||||
return std::make_unique<MIGraphXNotification>(*this);
|
||||
}
|
||||
|
||||
void MIGraphXStream::Flush() {
|
||||
if (own_stream_)
|
||||
HIP_CALL_THROW(hipStreamSynchronize(static_cast<hipStream_t>(GetHandle())));
|
||||
}
|
||||
|
||||
void MIGraphXStream::EnqueDeferredCPUBuffer(void* cpu_buffer) {
|
||||
// stream is per thread, so don't need lock
|
||||
deferred_cpu_buffers_.push_back(cpu_buffer);
|
||||
}
|
||||
|
||||
struct CpuBuffersInfo {
|
||||
// This struct stores the information needed
|
||||
// to release CPU buffers allocated for GPU kernels.
|
||||
// It's used to enqueue their release after
|
||||
// associated GPU kernels in a MIGraphX stream.
|
||||
|
||||
// This is a CPU allocator in MIGraphX EP.
|
||||
// It must be the one used to allocate the
|
||||
// following pointers.
|
||||
AllocatorPtr allocator;
|
||||
// buffers[i] is the i-th pointer added by
|
||||
// AddDeferredReleaseCPUPtr for a specific
|
||||
// MIGraphX stream. For example, this fields
|
||||
// should contain all values in
|
||||
// deferred_release_buffer_pool_[my_stream]
|
||||
// when release my_stream's buffers.
|
||||
std::unique_ptr<void*[]> buffers;
|
||||
// CPU buffer buffers[i].
|
||||
// Number of buffer points in "buffers".
|
||||
size_t n_buffers;
|
||||
};
|
||||
|
||||
static void ReleaseCpuBufferCallback(void* raw_info) {
|
||||
std::unique_ptr<CpuBuffersInfo> info = std::make_unique<CpuBuffersInfo>();
|
||||
info.reset(reinterpret_cast<CpuBuffersInfo*>(raw_info));
|
||||
for (size_t i = 0; i < info->n_buffers; ++i) {
|
||||
info->allocator->Free(info->buffers[i]);
|
||||
}
|
||||
}
|
||||
|
||||
Status MIGraphXStream::CleanUpOnRunEnd() {
|
||||
if (deferred_cpu_buffers_.empty())
|
||||
return Status::OK();
|
||||
// Release the ownership of cpu_buffers_info so that the underlying
|
||||
// object will keep alive until the end of ReleaseCpuBufferCallback.
|
||||
if (release_cpu_buffer_on_migraphx_stream_ && cpu_allocator_->Info().alloc_type == OrtArenaAllocator) {
|
||||
std::unique_ptr<CpuBuffersInfo> cpu_buffers_info = std::make_unique<CpuBuffersInfo>();
|
||||
cpu_buffers_info->allocator = cpu_allocator_;
|
||||
cpu_buffers_info->buffers = std::make_unique<void*[]>(deferred_cpu_buffers_.size());
|
||||
for (size_t i = 0; i < deferred_cpu_buffers_.size(); ++i) {
|
||||
cpu_buffers_info->buffers[i] = deferred_cpu_buffers_.at(i);
|
||||
}
|
||||
cpu_buffers_info->n_buffers = deferred_cpu_buffers_.size();
|
||||
HIP_RETURN_IF_ERROR(hipLaunchHostFunc(static_cast<hipStream_t>(GetHandle()), ReleaseCpuBufferCallback, cpu_buffers_info.release()));
|
||||
} else {
|
||||
HIP_RETURN_IF_ERROR(hipStreamSynchronize(static_cast<hipStream_t>(GetHandle())));
|
||||
for (auto* buffer : deferred_cpu_buffers_) {
|
||||
cpu_allocator_->Free(buffer);
|
||||
}
|
||||
}
|
||||
|
||||
deferred_cpu_buffers_.clear();
|
||||
return Status::OK();
|
||||
}
|
||||
|
||||
void* MIGraphXStream::GetResource(int version, int id) const {
|
||||
ORT_ENFORCE(version <= ORT_ROCM_RESOUCE_VERSION, "resource version unsupported!");
|
||||
void* resource{};
|
||||
switch (id) {
|
||||
case RocmResource::hip_stream_t:
|
||||
return reinterpret_cast<void*>(GetHandle());
|
||||
default:
|
||||
break;
|
||||
}
|
||||
return resource;
|
||||
}
|
||||
|
||||
// CPU Stream command handles
|
||||
void WaitMIGraphXNotificationOnDevice(Stream& stream, synchronize::Notification& notification) {
|
||||
static_cast<MIGraphXNotification*>(¬ification)->wait_on_device(stream);
|
||||
}
|
||||
|
||||
void WaitMIGraphXNotificationOnHost(Stream& /*stream*/, synchronize::Notification& notification) {
|
||||
static_cast<MIGraphXNotification*>(¬ification)->wait_on_host();
|
||||
}
|
||||
|
||||
void RegisterMIGraphXStreamHandles(IStreamCommandHandleRegistry& stream_handle_registry,
|
||||
const OrtDevice::DeviceType device_type,
|
||||
AllocatorPtr cpu_allocator,
|
||||
bool release_cpu_buffer_on_migraphx_stream,
|
||||
hipStream_t external_stream,
|
||||
bool use_existing_stream) {
|
||||
// wait migraphx notification on migraphx ep
|
||||
stream_handle_registry.RegisterWaitFn(device_type, device_type, WaitMIGraphXNotificationOnDevice);
|
||||
// wait migraphx notification on cpu ep
|
||||
stream_handle_registry.RegisterWaitFn(device_type, OrtDevice::CPU, WaitMIGraphXNotificationOnHost);
|
||||
if (!use_existing_stream)
|
||||
stream_handle_registry.RegisterCreateStreamFn(device_type, [cpu_allocator, release_cpu_buffer_on_migraphx_stream](const OrtDevice& device) {
|
||||
HIP_CALL_THROW(hipSetDevice(device.Id()));
|
||||
hipStream_t stream = nullptr;
|
||||
HIP_CALL_THROW(hipStreamCreateWithFlags(&stream, hipStreamNonBlocking));
|
||||
return std::make_unique<MIGraphXStream>(stream, device, cpu_allocator, release_cpu_buffer_on_migraphx_stream);
|
||||
});
|
||||
else
|
||||
stream_handle_registry.RegisterCreateStreamFn(device_type, [cpu_allocator,
|
||||
release_cpu_buffer_on_migraphx_stream,
|
||||
external_stream](const OrtDevice& device) {
|
||||
return std::make_unique<MIGraphXStream>(external_stream, device, cpu_allocator, release_cpu_buffer_on_migraphx_stream);
|
||||
});
|
||||
}
|
||||
|
||||
} // namespace onnxruntime
|
||||
|
|
@ -1,48 +0,0 @@
|
|||
// Copyright (c) Microsoft Corporation. All rights reserved.
|
||||
// Licensed under the MIT License.
|
||||
|
||||
#pragma once
|
||||
#include "core/framework/stream_handles.h"
|
||||
#include "migraphx_inc.h"
|
||||
#include "migraphx_call.h"
|
||||
|
||||
#define HIP_RETURN_IF_ERROR(expr) ORT_RETURN_IF_ERROR(HIP_CALL(expr))
|
||||
|
||||
namespace onnxruntime {
|
||||
void WaitMIGraphXNotificationOnDevice(Stream& stream, synchronize::Notification& notification);
|
||||
|
||||
struct MIGraphXStream : Stream {
|
||||
MIGraphXStream(hipStream_t stream,
|
||||
const OrtDevice& device,
|
||||
AllocatorPtr cpu_allocator,
|
||||
bool release_cpu_buffer_on_migraphx_stream);
|
||||
|
||||
~MIGraphXStream();
|
||||
|
||||
std::unique_ptr<synchronize::Notification> CreateNotification(size_t /*num_consumers*/) override;
|
||||
|
||||
void Flush() override;
|
||||
|
||||
Status CleanUpOnRunEnd() override;
|
||||
|
||||
void EnqueDeferredCPUBuffer(void* cpu_buffer);
|
||||
|
||||
bool own_stream_{true};
|
||||
|
||||
virtual void* GetResource(int version, int id) const;
|
||||
|
||||
virtual WaitNotificationFn GetWaitNotificationFn() const { return WaitMIGraphXNotificationOnDevice; }
|
||||
|
||||
private:
|
||||
std::vector<void*> deferred_cpu_buffers_;
|
||||
AllocatorPtr cpu_allocator_;
|
||||
bool release_cpu_buffer_on_migraphx_stream_{true};
|
||||
};
|
||||
|
||||
void RegisterMIGraphXStreamHandles(IStreamCommandHandleRegistry& stream_handle_registry,
|
||||
const OrtDevice::DeviceType device_type,
|
||||
AllocatorPtr cpu_allocator,
|
||||
bool release_cpu_buffer_on_migraphx_stream,
|
||||
hipStream_t external_stream,
|
||||
bool use_existing_stream);
|
||||
} // namespace onnxruntime
|
||||
|
|
@ -279,9 +279,6 @@ std::unique_ptr<IAllocator> CreateCPUAllocator(const OrtMemoryInfo& memory_info)
|
|||
std::unique_ptr<IAllocator> CreateCUDAAllocator(int16_t device_id, const char* name);
|
||||
std::unique_ptr<IAllocator> CreateCUDAPinnedAllocator(const char* name);
|
||||
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXAllocator(int16_t device_id, const char* name);
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXPinnedAllocator(int16_t device_id, const char* name);
|
||||
|
||||
std::unique_ptr<IAllocator> CreateROCMAllocator(int16_t device_id, const char* name);
|
||||
std::unique_ptr<IAllocator> CreateROCMPinnedAllocator(const char* name);
|
||||
|
||||
|
|
|
|||
|
|
@ -353,12 +353,16 @@ std::unique_ptr<IDataTransfer> CreateGPUDataTransfer() {
|
|||
#endif
|
||||
|
||||
#ifdef USE_MIGRAPHX
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXAllocator(int16_t device_id, const char* name) {
|
||||
return g_host->CreateMIGraphXAllocator(device_id, name);
|
||||
std::unique_ptr<IAllocator> CreateROCMAllocator(int16_t device_id, const char* name) {
|
||||
return g_host->CreateROCMAllocator(device_id, name);
|
||||
}
|
||||
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXPinnedAllocator(int16_t device_id, const char* name) {
|
||||
return g_host->CreateMIGraphXPinnedAllocator(device_id, name);
|
||||
std::unique_ptr<IAllocator> CreateROCMPinnedAllocator(const char* name) {
|
||||
return g_host->CreateROCMPinnedAllocator(name);
|
||||
}
|
||||
|
||||
std::unique_ptr<IDataTransfer> CreateGPUDataTransfer() {
|
||||
return g_host->CreateGPUDataTransfer();
|
||||
}
|
||||
#endif
|
||||
|
||||
|
|
|
|||
|
|
@ -178,11 +178,6 @@ struct ProviderHost {
|
|||
virtual void CudaCall_true(int retCode, const char* exprString, const char* libName, int successCode, const char* msg, const char* file, const int line) = 0;
|
||||
#endif
|
||||
|
||||
#ifdef USE_MIGRAPHX
|
||||
virtual std::unique_ptr<IAllocator> CreateMIGraphXAllocator(int16_t device_id, const char* name) = 0;
|
||||
virtual std::unique_ptr<IAllocator> CreateMIGraphXPinnedAllocator(int16_t device_id, const char* name) = 0;
|
||||
#endif
|
||||
|
||||
#ifdef USE_ROCM
|
||||
virtual std::unique_ptr<IAllocator> CreateROCMAllocator(int16_t device_id, const char* name) = 0;
|
||||
virtual std::unique_ptr<IAllocator> CreateROCMPinnedAllocator(const char* name) = 0;
|
||||
|
|
|
|||
|
|
@ -130,8 +130,6 @@ ProviderInfo_Dnnl& GetProviderInfo_Dnnl();
|
|||
ProviderInfo_ROCM* TryGetProviderInfo_ROCM();
|
||||
ProviderInfo_ROCM& GetProviderInfo_ROCM();
|
||||
ProviderHostCPU& GetProviderHostCPU();
|
||||
ProviderInfo_MIGraphX* TryGetProviderInfo_MIGraphX();
|
||||
ProviderInfo_MIGraphX& GetProviderInfo_MIGraphX();
|
||||
ONNX_NAMESPACE::OpSchema CreateSchema(const std::string& domain, const std::vector<const OrtCustomOp*>& ops);
|
||||
struct TensorShapeProto_Dimension_Iterator_Impl : TensorShapeProto_Dimension_Iterator {
|
||||
TensorShapeProto_Dimension_Iterator_Impl(google::protobuf::internal::RepeatedPtrIterator<const onnx::TensorShapeProto_Dimension>&& v) : v_{std::move(v)} {}
|
||||
|
|
@ -243,11 +241,6 @@ struct ProviderHostImpl : ProviderHost {
|
|||
void CudaCall_true(int retCode, const char* exprString, const char* libName, int successCode, const char* msg, const char* file, const int line) override { GetProviderInfo_CUDA().CudaCall_true(retCode, exprString, libName, successCode, msg, file, line); }
|
||||
#endif
|
||||
|
||||
#ifdef USE_MIGRAPHX
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXAllocator(int16_t device_id, const char* name) override { return GetProviderInfo_MIGraphX().CreateMIGraphXAllocator(device_id, name); }
|
||||
std::unique_ptr<IAllocator> CreateMIGraphXPinnedAllocator(int16_t device_id, const char* name) override { return GetProviderInfo_MIGraphX().CreateMIGraphXPinnedAllocator(device_id, name); }
|
||||
#endif
|
||||
|
||||
#ifdef USE_ROCM
|
||||
std::unique_ptr<IAllocator> CreateROCMAllocator(int16_t device_id, const char* name) override { return GetProviderInfo_ROCM().CreateROCMAllocator(device_id, name); }
|
||||
std::unique_ptr<IAllocator> CreateROCMPinnedAllocator(const char* name) override { return GetProviderInfo_ROCM().CreateROCMPinnedAllocator(name); }
|
||||
|
|
@ -1907,20 +1900,6 @@ ProviderInfo_ROCM& GetProviderInfo_ROCM() {
|
|||
ORT_THROW("ROCM Provider not available, can't get interface for it");
|
||||
}
|
||||
|
||||
ProviderInfo_MIGraphX* TryGetProviderInfo_MIGraphX() try {
|
||||
return reinterpret_cast<ProviderInfo_MIGraphX*>(s_library_migraphx.Get().GetInfo());
|
||||
} catch (const std::exception& exception) {
|
||||
LOGS_DEFAULT(ERROR) << exception.what();
|
||||
return nullptr;
|
||||
}
|
||||
|
||||
ProviderInfo_MIGraphX& GetProviderInfo_MIGraphX() {
|
||||
if (auto* info = TryGetProviderInfo_MIGraphX())
|
||||
return *info;
|
||||
|
||||
ORT_THROW("MIGraphX Provider not available, can't get interface for it");
|
||||
}
|
||||
|
||||
void CopyGpuToCpu(
|
||||
void* dst_ptr,
|
||||
const void* src_ptr,
|
||||
|
|
|
|||
63
setup.py
63
setup.py
|
|
@ -56,7 +56,6 @@ wheel_name_suffix = parse_arg_remove_string(sys.argv, "--wheel_name_suffix=")
|
|||
|
||||
cuda_version = None
|
||||
rocm_version = None
|
||||
is_migraphx = False
|
||||
is_rocm = False
|
||||
is_openvino = False
|
||||
# The following arguments are mutually exclusive
|
||||
|
|
@ -65,9 +64,8 @@ if wheel_name_suffix == "gpu":
|
|||
cuda_version = parse_arg_remove_string(sys.argv, "--cuda_version=")
|
||||
elif parse_arg_remove_boolean(sys.argv, "--use_rocm"):
|
||||
is_rocm = True
|
||||
package_name = "onnxruntime-rocm" if not nightly_build else "ort-rocm-nightly"
|
||||
rocm_version = parse_arg_remove_string(sys.argv, "--rocm_version=")
|
||||
elif parse_arg_remove_boolean(sys.argv, "--use_migraphx"):
|
||||
is_migraphx = True
|
||||
elif parse_arg_remove_boolean(sys.argv, "--use_openvino"):
|
||||
is_openvino = True
|
||||
package_name = "onnxruntime-openvino"
|
||||
|
|
@ -89,9 +87,6 @@ elif parse_arg_remove_boolean(sys.argv, "--use_azure"):
|
|||
elif parse_arg_remove_boolean(sys.argv, "--use_qnn"):
|
||||
package_name = "onnxruntime-qnn"
|
||||
|
||||
if is_rocm or is_migraphx:
|
||||
package_name = "onnxruntime-rocm" if not nightly_build else "ort-rocm-nightly"
|
||||
|
||||
# PEP 513 defined manylinux1_x86_64 and manylinux1_i686
|
||||
# PEP 571 defined manylinux2010_x86_64 and manylinux2010_i686
|
||||
# PEP 599 defines the following platform tags:
|
||||
|
|
@ -285,21 +280,10 @@ class InstallCommand(InstallCommandBase):
|
|||
return ret
|
||||
|
||||
|
||||
providers_cuda_or_rocm = "onnxruntime_providers_" + ("rocm" if is_rocm else "cuda")
|
||||
providers_tensorrt_or_migraphx = "onnxruntime_providers_" + ("migraphx" if is_migraphx else "tensorrt")
|
||||
providers_openvino = "onnxruntime_providers_openvino"
|
||||
providers_cann = "onnxruntime_providers_cann"
|
||||
|
||||
if platform.system() == "Linux":
|
||||
providers_cuda_or_rocm = "lib" + providers_cuda_or_rocm + ".so"
|
||||
providers_tensorrt_or_migraphx = "lib" + providers_tensorrt_or_migraphx + ".so"
|
||||
providers_openvino = "lib" + providers_openvino + ".so"
|
||||
providers_cann = "lib" + providers_cann + ".so"
|
||||
elif platform.system() == "Windows":
|
||||
providers_cuda_or_rocm = providers_cuda_or_rocm + ".dll"
|
||||
providers_tensorrt_or_migraphx = providers_tensorrt_or_migraphx + ".dll"
|
||||
providers_openvino = providers_openvino + ".dll"
|
||||
providers_cann = providers_cann + ".dll"
|
||||
providers_cuda_or_rocm = "libonnxruntime_providers_" + ("rocm.so" if is_rocm else "cuda.so")
|
||||
providers_tensorrt_or_migraphx = "libonnxruntime_providers_" + ("migraphx.so" if is_rocm else "tensorrt.so")
|
||||
providers_openvino = "libonnxruntime_providers_openvino.so"
|
||||
providers_cann = "libonnxruntime_providers_cann.so"
|
||||
|
||||
# Additional binaries
|
||||
dl_libs = []
|
||||
|
|
@ -313,22 +297,19 @@ if platform.system() == "Linux":
|
|||
"libmklml_gnu.so",
|
||||
"libiomp5.so",
|
||||
"mimalloc.so",
|
||||
# DNNL, TensorRT & OpenVINO EPs are built as shared libs
|
||||
"libonnxruntime_providers_shared.so",
|
||||
"libonnxruntime_providers_dnnl.so",
|
||||
"libonnxruntime_providers_openvino.so",
|
||||
"libonnxruntime_providers_vitisai.so",
|
||||
providers_cuda_or_rocm,
|
||||
providers_tensorrt_or_migraphx,
|
||||
providers_cann,
|
||||
]
|
||||
dl_libs = [
|
||||
"libonnxruntime_providers_shared.so",
|
||||
providers_cuda_or_rocm,
|
||||
providers_tensorrt_or_migraphx,
|
||||
providers_cann,
|
||||
]
|
||||
|
||||
dl_libs = ["libonnxruntime_providers_shared.so"]
|
||||
dl_libs.append(providers_cuda_or_rocm)
|
||||
dl_libs.append(providers_tensorrt_or_migraphx)
|
||||
dl_libs.append(providers_cann)
|
||||
# DNNL, TensorRT & OpenVINO EPs are built as shared libs
|
||||
libs.extend(["libonnxruntime_providers_shared.so"])
|
||||
libs.extend(["libonnxruntime_providers_dnnl.so"])
|
||||
libs.extend(["libonnxruntime_providers_openvino.so"])
|
||||
libs.extend(["libonnxruntime_providers_vitisai.so"])
|
||||
libs.append(providers_cuda_or_rocm)
|
||||
libs.append(providers_tensorrt_or_migraphx)
|
||||
libs.append(providers_cann)
|
||||
if nightly_build:
|
||||
libs.extend(["libonnxruntime_pywrapper.so"])
|
||||
elif platform.system() == "Darwin":
|
||||
|
|
@ -342,15 +323,7 @@ elif platform.system() == "Darwin":
|
|||
if nightly_build:
|
||||
libs.extend(["libonnxruntime_pywrapper.dylib"])
|
||||
else:
|
||||
libs = [
|
||||
"onnxruntime_pybind11_state.pyd",
|
||||
"dnnl.dll",
|
||||
"mklml.dll",
|
||||
"libiomp5md.dll",
|
||||
providers_cuda_or_rocm,
|
||||
providers_tensorrt_or_migraphx,
|
||||
providers_cann,
|
||||
]
|
||||
libs = ["onnxruntime_pybind11_state.pyd", "dnnl.dll", "mklml.dll", "libiomp5md.dll"]
|
||||
# DNNL, TensorRT & OpenVINO EPs are built as shared libs
|
||||
libs.extend(["onnxruntime_providers_shared.dll"])
|
||||
libs.extend(["onnxruntime_providers_dnnl.dll"])
|
||||
|
|
|
|||
|
|
@ -611,7 +611,6 @@ def parse_arguments():
|
|||
"MinGW Makefiles",
|
||||
"Ninja",
|
||||
"NMake Makefiles",
|
||||
"NMake Makefiles JOM",
|
||||
"Unix Makefiles",
|
||||
"Visual Studio 17 2022",
|
||||
"Xcode",
|
||||
|
|
@ -2208,7 +2207,6 @@ def build_python_wheel(
|
|||
use_cuda,
|
||||
cuda_version,
|
||||
use_rocm,
|
||||
use_migraphx,
|
||||
rocm_version,
|
||||
use_dnnl,
|
||||
use_tensorrt,
|
||||
|
|
@ -2260,8 +2258,6 @@ def build_python_wheel(
|
|||
args.append("--use_rocm")
|
||||
if rocm_version:
|
||||
args.append(f"--rocm_version={rocm_version}")
|
||||
elif use_migraphx:
|
||||
args.append("--use_migraphx")
|
||||
elif use_openvino:
|
||||
args.append("--use_openvino")
|
||||
elif use_dnnl:
|
||||
|
|
@ -2587,6 +2583,9 @@ def main():
|
|||
if args.use_tensorrt:
|
||||
args.use_cuda = True
|
||||
|
||||
if args.use_migraphx:
|
||||
args.use_rocm = True
|
||||
|
||||
if args.build_wheel or args.gen_doc or args.use_tvm or args.enable_training:
|
||||
args.enable_pybind = True
|
||||
|
||||
|
|
@ -2873,8 +2872,7 @@ def main():
|
|||
# fail unexpectedly. Similar, if your packaging step forgot to copy a file into the package, we don't know it
|
||||
# either.
|
||||
if args.build:
|
||||
# TODO: find asan DLL and copy it to onnxruntime/capi folder when args.enable_address_sanitizer is True and
|
||||
# the target OS is Windows
|
||||
# TODO: find asan DLL and copy it to onnxruntime/capi folder when args.enable_address_sanitizer is True and the target OS is Windows
|
||||
if args.build_wheel:
|
||||
nightly_build = bool(os.getenv("NIGHTLY_BUILD") == "1")
|
||||
default_training_package_device = bool(os.getenv("DEFAULT_TRAINING_PACKAGE_DEVICE") == "1")
|
||||
|
|
@ -2885,7 +2883,6 @@ def main():
|
|||
args.use_cuda,
|
||||
args.cuda_version,
|
||||
args.use_rocm,
|
||||
args.use_migraphx,
|
||||
args.rocm_version,
|
||||
args.use_dnnl,
|
||||
args.use_tensorrt,
|
||||
|
|
|
|||
Loading…
Reference in a new issue