AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

SWDEV-332606 - SWDEV-333585 - Fix hipMemcpySymbol* and hipGetSymbol* APIs

Browse Source 
This also fixes the QMCPack staging (SWDEV-335826) test failures related
to hipMemcpyToSymbol

Change-Id: Icc3ff04fb1e17999f5bddf12e1a7266e2289032f
This commit is contained in:
Satyanvesh Dittakavi
2022-05-12 08:52:27 +00:00
parent 4561a207dd
commit 8a333a989b
2 changed files with 24 additions and 0 deletions
Download Patch File Download Diff File Expand all files Collapse all files
hipamd/src/hip_memory.cpp
+18
View File
@@ -1125,6 +1125,11 @@ inline hipError_t ihipMemcpySymbol_validate(const void* symbol, size_t sizeBytes
hipError_t hipMemcpyToSymbol_common(const void* symbol, const void* src, size_t sizeBytes,
size_t offset, hipMemcpyKind kind, hipStream_t stream=nullptr) {
CHECK_STREAM_CAPTURING();
if (kind != hipMemcpyHostToDevice && kind != hipMemcpyDeviceToDevice) {
HIP_RETURN(hipErrorInvalidMemcpyDirection);
}
size_t sym_size = 0;
hipDeviceptr_t device_ptr = nullptr;
@@ -1153,6 +1158,11 @@ hipError_t hipMemcpyToSymbol_spt(const void* symbol, const void* src, size_t siz
hipError_t hipMemcpyFromSymbol_common(void* dst, const void* symbol, size_t sizeBytes,
size_t offset, hipMemcpyKind kind, hipStream_t stream=nullptr) {
CHECK_STREAM_CAPTURING();
if (kind != hipMemcpyDeviceToHost && kind != hipMemcpyDeviceToDevice) {
HIP_RETURN(hipErrorInvalidMemcpyDirection);
}
size_t sym_size = 0;
hipDeviceptr_t device_ptr = nullptr;
@@ -1184,6 +1194,10 @@ hipError_t hipMemcpyToSymbolAsync(const void* symbol, const void* src, size_t si
STREAM_CAPTURE(hipMemcpyToSymbolAsync, stream, symbol, src, sizeBytes, offset, kind);
if (kind != hipMemcpyHostToDevice && kind != hipMemcpyDeviceToDevice) {
HIP_RETURN(hipErrorInvalidMemcpyDirection);
}
size_t sym_size = 0;
hipDeviceptr_t device_ptr = nullptr;
@@ -1201,6 +1215,10 @@ hipError_t hipMemcpyFromSymbolAsync(void* dst, const void* symbol, size_t sizeBy
STREAM_CAPTURE(hipMemcpyFromSymbolAsync, stream, dst, symbol, sizeBytes, offset, kind);
if (kind != hipMemcpyDeviceToHost && kind != hipMemcpyDeviceToDevice) {
HIP_RETURN(hipErrorInvalidMemcpyDirection);
}
size_t sym_size = 0;
hipDeviceptr_t device_ptr = nullptr;
hipamd/src/hip_platform.cpp
+6
View File
@@ -278,6 +278,9 @@ hipError_t hipGetSymbolAddress(void** devPtr, const void* symbol) {
HIP_INIT_API(hipGetSymbolAddress, devPtr, symbol);
hipError_t hip_error = hipSuccess;
if (devPtr == nullptr) {
HIP_RETURN(hipErrorInvalidValue);
}
size_t sym_size = 0;
HIP_RETURN_ONFAIL(PlatformState::instance().getStatGlobalVar(symbol, ihipGetDevice(), devPtr, &sym_size));
@@ -288,6 +291,9 @@ hipError_t hipGetSymbolAddress(void** devPtr, const void* symbol) {
hipError_t hipGetSymbolSize(size_t* sizePtr, const void* symbol) {
HIP_INIT_API(hipGetSymbolSize, sizePtr, symbol);
if (sizePtr == nullptr) {
HIP_RETURN(hipErrorInvalidValue);
}
hipDeviceptr_t device_ptr = nullptr;
HIP_RETURN_ONFAIL(PlatformState::instance().getStatGlobalVar(symbol, ihipGetDevice(), &device_ptr, sizePtr));