diff --git a/hipamd/src/hip_context.cpp b/hipamd/src/hip_context.cpp index f58768792f..32b3545ecd 100644 --- a/hipamd/src/hip_context.cpp +++ b/hipamd/src/hip_context.cpp @@ -132,6 +132,9 @@ int getDeviceID(amd::Context& ctx) { // ================================================================================================ hip::Stream* getNullStream(bool wait ) { Device* device = getCurrentDevice(); + if (device == nullptr) { + LogError("Invalid device"); + } return device ? device->NullStream(wait) : nullptr; } diff --git a/hipamd/src/hip_device.cpp b/hipamd/src/hip_device.cpp index 4d2d408abb..a05c7b9bb5 100644 --- a/hipamd/src/hip_device.cpp +++ b/hipamd/src/hip_device.cpp @@ -39,6 +39,7 @@ hip::Stream* Device::NullStream(bool wait) { } } if (null_stream_ == nullptr) { + LogError("Cannot create new Stream object"); return nullptr; } if (wait == true) { diff --git a/hipamd/src/hip_device_runtime.cpp b/hipamd/src/hip_device_runtime.cpp index 0d2d2d6cf3..bf9196461a 100644 --- a/hipamd/src/hip_device_runtime.cpp +++ b/hipamd/src/hip_device_runtime.cpp @@ -657,7 +657,7 @@ hipError_t hipGetDevice(int* deviceId) { HIP_RETURN(hipErrorNoDevice); } *deviceId = dev; - HIP_RETURN(hipSuccess); + HIP_RETURN(hipSuccess, *deviceId); } else { HIP_RETURN(hipErrorInvalidValue); } diff --git a/hipamd/src/hip_memory.cpp b/hipamd/src/hip_memory.cpp index 482916bd0a..84ddee4c13 100644 --- a/hipamd/src/hip_memory.cpp +++ b/hipamd/src/hip_memory.cpp @@ -1360,7 +1360,24 @@ hipError_t hipHostAlloc(void** ptr, size_t sizeBytes, unsigned int flags) { hipError_t status = ihipHostMalloc(ptr, sizeBytes, flags); HIP_RETURN_DURATION(status, *ptr); -}; +} + +hipError_t hipMemcpyAsync_common(void* dst, const void* src, size_t sizeBytes, + hipMemcpyKind kind, hipStream_t stream) { + STREAM_CAPTURE(hipMemcpyAsync, stream, dst, src, sizeBytes, kind); + + if (static_cast(kind) > hipMemcpyDefault && kind != hipMemcpyDeviceToDeviceNoCU) { + return hipErrorInvalidMemcpyDirection; + } + hip::Stream* hip_stream = hip::getStream(stream); + if (hip_stream == nullptr) { + return hipErrorInvalidValue; + } + if (!hip::isValid(stream)) { + return hipErrorContextIsDestroyed; + } + return ihipMemcpy(dst, src, sizeBytes, kind, *hip_stream, true); +} inline hipError_t ihipMemcpySymbol_validate(const void* symbol, size_t sizeBytes, size_t offset, size_t &sym_size, hipDeviceptr_t &device_ptr) { @@ -1462,8 +1479,8 @@ hipError_t hipMemcpyToSymbolAsync_common(const void* symbol, const void* src, si if (status != hipSuccess) { return status; } - /* Copy memory from source to destination address */ - return hipMemcpyAsync(device_ptr, src, sizeBytes, kind, stream); + + return hipMemcpyAsync_common(device_ptr, src, sizeBytes, kind, stream); } hipError_t hipMemcpyToSymbolAsync(const void* symbol, const void* src, size_t sizeBytes, @@ -1549,23 +1566,6 @@ hipError_t hipMemcpyDtoD(hipDeviceptr_t dstDevice, HIP_RETURN_DURATION(ihipMemcpy(dstDevice, srcDevice, ByteCount, hipMemcpyDeviceToDevice, *stream)); } -hipError_t hipMemcpyAsync_common(void* dst, const void* src, size_t sizeBytes, - hipMemcpyKind kind, hipStream_t stream) { - STREAM_CAPTURE(hipMemcpyAsync, stream, dst, src, sizeBytes, kind); - - if (static_cast(kind) > hipMemcpyDefault && kind != hipMemcpyDeviceToDeviceNoCU) { - return hipErrorInvalidMemcpyDirection; - } - hip::Stream* hip_stream = hip::getStream(stream); - if (hip_stream == nullptr) { - return hipErrorInvalidValue; - } - if (!hip::isValid(stream)) { - return hipErrorContextIsDestroyed; - } - return ihipMemcpy(dst, src, sizeBytes, kind, *hip_stream, true); -} - hipError_t hipMemcpyAsync(void* dst, const void* src, size_t sizeBytes, hipMemcpyKind kind, hipStream_t stream) { HIP_INIT_API(hipMemcpyAsync, dst, src, sizeBytes, kind, stream);