saymrwulf/onnxruntime
1
0
Fork 0
mirror of https://github.com/saymrwulf/onnxruntime.git synced 2026-06-04 23:59:56 +00:00
Code Issues Projects Releases Packages Wiki Activity Actions

[MigraphX] Fix potential synchronization problem when ORT_ENABLE_STREAM is true (#22589)

Browse source 
### Description
Replace `hipMemcpy` with `hipMemcpyWithStream`



### Motivation and Context
`hipMemcpy` uses default stream, which may be out of synchronization
with the current stream when ORT_ENABLE_STREAM is defined.
This commit is contained in:
Xinya Zhang 2024-10-25 13:19:59 -05:00 committed by GitHub
parent 7acbd51912
commit 28efacfd5a
No known key found for this signature in database
GPG key ID: B5690EEEBB952194
2 changed files with 6 additions and 2 deletions
Show stats Download patch file Download diff file Expand all files Collapse all files

2
onnxruntime/core/providers/migraphx/gpu_data_transfer.cc
View file

@ -57,7 +57,7 @@ common::Status GPUDataTransfer::CopyTensorAsync(const Tensor& src, Tensor& dst,
HIP_CALL_THROW(hipMemcpyAsync(dst_data, src_data, bytes, hipMemcpyDeviceToDevice, static_cast<hipStream_t>(stream.GetHandle())));
} else {
// copy from other CPU memory to GPU, this is blocking
HIP_CALL_THROW(hipMemcpy(dst_data, src_data, bytes, hipMemcpyHostToDevice));
HIP_CALL_THROW(hipMemcpyWithStream(dst_data, src_data, bytes, hipMemcpyHostToDevice, static_cast<hipStream_t>(stream.GetHandle())));
}
} else if (src_device.Type() == OrtDevice::GPU) {
HIP_CALL_THROW(hipMemcpyAsync(dst_data, src_data, bytes, hipMemcpyDeviceToHost, static_cast<hipStream_t>(stream.GetHandle())));

6
onnxruntime/core/providers/migraphx/migraphx_execution_provider.cc
View file

@ -1445,7 +1445,11 @@ Status MIGraphXExecutionProvider::Compile(const std::vector<FusedNodeAndGraph>&
std::vector<int64_t> ort_shape{res_lens.begin(), res_lens.end()};
auto output_tensor = ctx.GetOutput(i, ort_shape.data(), ort_shape.size());
void* output_data = output_tensor.GetTensorMutableRawData();
HIP_CALL_THROW(hipMemcpy(output_data, gpu_res.data(), res_shape.bytes(), hipMemcpyDeviceToDevice));
HIP_CALL_THROW(hipMemcpyWithStream(output_data,
gpu_res.data(),
res_shape.bytes(),
hipMemcpyDeviceToDevice,
static_cast<hipStream_t>(rocm_stream)));
}
}
};