From 883fdfb820a52e1352b150519d8b046d25c45d0d Mon Sep 17 00:00:00 2001 From: Sourabh U Betigeri Date: Thu, 18 Dec 2025 15:40:13 -0800 Subject: [PATCH] Revert "clr: Minor fixes for error return" (#2399) - This reverts commit 8dd8436e43c7f0d062fd73252bf61c35615d181d. - Resolve MIOpen test failures observed in TheRock - TheRock Issue: ROCm/TheRock#2642 - room-systems issue: #2400 --- projects/clr/hipamd/src/hip_event.cpp | 8 +++--- projects/clr/hipamd/src/hip_memory.cpp | 27 ++++++++------------ projects/clr/hipamd/src/hip_stream.cpp | 35 +++++++++++--------------- 3 files changed, 28 insertions(+), 42 deletions(-) diff --git a/projects/clr/hipamd/src/hip_event.cpp b/projects/clr/hipamd/src/hip_event.cpp index 6dfbfc9919..3a97124899 100644 --- a/projects/clr/hipamd/src/hip_event.cpp +++ b/projects/clr/hipamd/src/hip_event.cpp @@ -297,14 +297,14 @@ hipError_t ihipEventCreateWithFlags(hipEvent_t* event, unsigned flags) { e = new hip::Event(flags); } } - if (e == nullptr) { - return hipErrorOutOfMemory; - } // App might have used combination of flags i.e. hipEventInterprocess|hipEventDisableTiming // However based on hipEventInterprocess flag, IPCEvent creates even with // JUST hipEventInterprocess and hence, Actual hipEventInterprocess|hipEventDisableTiming // flag is getting supressed with hipEventInterprocess e->flags_ = flags; + if (e == nullptr) { + return hipErrorOutOfMemory; + } *event = reinterpret_cast(e); std::unique_lock lock(hip::eventSetLock); hip::eventSet.insert(*event); @@ -346,7 +346,7 @@ hipError_t hipEventDestroy(hipEvent_t event) { std::unique_lock lock(hip::eventSetLock); if (hip::eventSet.erase(event) == 0) { - HIP_RETURN(hipErrorContextIsDestroyed); + return hipErrorContextIsDestroyed; } hip::Event* e = reinterpret_cast(event); diff --git a/projects/clr/hipamd/src/hip_memory.cpp b/projects/clr/hipamd/src/hip_memory.cpp index 7275de5fd3..f802c4b68c 100644 --- a/projects/clr/hipamd/src/hip_memory.cpp +++ b/projects/clr/hipamd/src/hip_memory.cpp @@ -80,6 +80,7 @@ amd::Memory* getMemoryObjectWithOffset(const void* ptr, const size_t size) { } memObj = new (memObj->getContext()) amd::Buffer(*memObj, memObj->getMemFlags(), offset, size); if (memObj == nullptr) { + ; return nullptr; } @@ -271,7 +272,7 @@ hipError_t hipSignalExternalSemaphoresAsync(const hipExternalSemaphore_t* extSem *hip_stream, extSemArray[i], paramsArray[i].params.fence.value, amd::ExternalSemaphoreCmd::COMMAND_SIGNAL_EXTSEMAPHORE); if (command == nullptr) { - HIP_RETURN(hipErrorOutOfMemory); + return hipErrorOutOfMemory; } command->enqueue(); command->release(); @@ -309,7 +310,7 @@ hipError_t hipWaitExternalSemaphoresAsync(const hipExternalSemaphore_t* extSemAr *hip_stream, extSemArray[i], paramsArray[i].params.fence.value, amd::ExternalSemaphoreCmd::COMMAND_WAIT_EXTSEMAPHORE); if (command == nullptr) { - HIP_RETURN(hipErrorOutOfMemory); + return hipErrorOutOfMemory; } command->enqueue(); command->release(); @@ -846,10 +847,6 @@ hipError_t hipMemcpyWithStream(void* dst, const void* src, size_t sizeBytes, hip hipError_t hipMemPtrGetInfo(void* ptr, size_t* size) { HIP_INIT_API(hipMemPtrGetInfo, ptr, size); - if (size == nullptr) { - HIP_RETURN(hipErrorInvalidValue); - } - if (ptr == nullptr) { *size = 0; HIP_RETURN(hipSuccess); @@ -915,10 +912,6 @@ hipError_t hipFreeArray(hipArray_t array) { hipError_t hipMemGetAddressRange(hipDeviceptr_t* pbase, size_t* psize, hipDeviceptr_t dptr) { HIP_INIT_API(hipMemGetAddressRange, pbase, psize, dptr); - if (pbase == nullptr || psize == nullptr) { - HIP_RETURN(hipErrorInvalidValue); - } - // Since we are using SVM buffer DevicePtr and HostPtr is the same void* ptr = dptr; size_t offset = 0; @@ -1224,7 +1217,7 @@ hipError_t ihipArrayCreate(hipArray_t* array, const HIP_ARRAY3D_DESCRIPTOR* pAll hipError_t hipArrayCreate(hipArray_t* array, const HIP_ARRAY_DESCRIPTOR* pAllocateArray) { HIP_INIT_API(hipArrayCreate, array, pAllocateArray); if (pAllocateArray == nullptr) { - HIP_RETURN(hipErrorInvalidValue); + return hipErrorInvalidValue; } CHECK_STREAM_CAPTURE_SUPPORTED(); HIP_ARRAY3D_DESCRIPTOR desc = { @@ -1639,10 +1632,10 @@ hipError_t hipMemcpyHtoDAsync(hipDeviceptr_t dstDevice, const void* srcHost, siz hipMemcpyKind kind = hipMemcpyHostToDevice; STREAM_CAPTURE(hipMemcpyHtoDAsync, stream, dstDevice, srcHost, ByteCount, kind); if (static_cast(kind) > hipMemcpyDefault && kind != hipMemcpyDeviceToDeviceNoCU) { - HIP_RETURN(hipErrorInvalidMemcpyDirection); + return hipErrorInvalidMemcpyDirection; } if (!hip::isValid(stream)) { - HIP_RETURN(hipErrorContextIsDestroyed); + return hipErrorContextIsDestroyed; } hip::Stream* hip_stream = hip::getStream(stream); if (hip_stream == nullptr) { @@ -1657,10 +1650,10 @@ hipError_t hipMemcpyDtoDAsync(hipDeviceptr_t dstDevice, hipDeviceptr_t srcDevice hipMemcpyKind kind = hipMemcpyDeviceToDevice; STREAM_CAPTURE(hipMemcpyDtoDAsync, stream, dstDevice, srcDevice, ByteCount, kind); if (static_cast(kind) > hipMemcpyDefault && kind != hipMemcpyDeviceToDeviceNoCU) { - HIP_RETURN(hipErrorInvalidMemcpyDirection); + return hipErrorInvalidMemcpyDirection; } if (!hip::isValid(stream)) { - HIP_RETURN(hipErrorContextIsDestroyed); + return hipErrorContextIsDestroyed; } hip::Stream* hip_stream = hip::getStream(stream); if (hip_stream == nullptr) { @@ -1675,10 +1668,10 @@ hipError_t hipMemcpyDtoHAsync(void* dstHost, hipDeviceptr_t srcDevice, size_t By hipMemcpyKind kind = hipMemcpyDeviceToHost; STREAM_CAPTURE(hipMemcpyDtoHAsync, stream, dstHost, srcDevice, ByteCount, kind); if (static_cast(kind) > hipMemcpyDefault && kind != hipMemcpyDeviceToDeviceNoCU) { - HIP_RETURN(hipErrorInvalidMemcpyDirection); + return hipErrorInvalidMemcpyDirection; } if (!hip::isValid(stream)) { - HIP_RETURN(hipErrorContextIsDestroyed); + return hipErrorContextIsDestroyed; } hip::Stream* hip_stream = hip::getStream(stream); if (hip_stream == nullptr) { diff --git a/projects/clr/hipamd/src/hip_stream.cpp b/projects/clr/hipamd/src/hip_stream.cpp index 35045d6ae0..bb087adff0 100644 --- a/projects/clr/hipamd/src/hip_stream.cpp +++ b/projects/clr/hipamd/src/hip_stream.cpp @@ -331,17 +331,13 @@ hipError_t hipDeviceGetStreamPriorityRange(int* leastPriority, int* greatestPrio // ================================================================================================ hipError_t hipStreamGetFlags_common(hipStream_t stream, unsigned int* flags) { - if (flags == nullptr || stream == nullptr) { + if ((flags != nullptr) && (stream != nullptr)) { + getStreamPerThread(stream); + *flags = reinterpret_cast(stream)->Flags(); + } else { return hipErrorInvalidValue; } - getStreamPerThread(stream); - - if (!hip::isValid(stream)) { - return hipErrorInvalidResourceHandle; - } - - *flags = reinterpret_cast(stream)->Flags(); return hipSuccess; } @@ -733,22 +729,18 @@ hipError_t hipExtStreamCreateWithCUMask(hipStream_t* stream, uint32_t cuMaskSize // ================================================================================================ hipError_t hipStreamGetPriority_common(hipStream_t stream, int* priority) { - if (priority == nullptr) { - return hipErrorInvalidValue; - } - - if (stream == nullptr) { + if ((priority != nullptr) && (stream == nullptr)) { *priority = 0; return hipSuccess; } - getStreamPerThread(stream); - - if (!hip::isValid(stream)) { - return hipErrorInvalidResourceHandle; + if ((priority != nullptr) && (stream != nullptr)) { + getStreamPerThread(stream); + *priority = static_cast(reinterpret_cast(stream)->GetPriority()); + } else { + return hipErrorInvalidValue; } - *priority = static_cast(reinterpret_cast(stream)->GetPriority()); return hipSuccess; } @@ -792,13 +784,14 @@ hipError_t hipExtStreamGetCUMask(hipStream_t stream, uint32_t cuMaskSize, uint32 uint32_t temp = 0; uint32_t bit_index = 0; for (uint32_t i = 0; i < info.maxComputeUnits_; i++) { - temp |= 1U << bit_index; - bit_index += 1; + temp |= 1UL << bit_index; if (bit_index >= 32) { defaultCUMask.push_back(temp); temp = 0; bit_index = 0; + temp |= 1UL << bit_index; } + bit_index += 1; } if (bit_index != 0) { defaultCUMask.push_back(temp); @@ -916,7 +909,7 @@ hipError_t hipStreamGetAttribute(hipStream_t stream, hipStreamAttrID attr, HIP_INIT_API(hipStreamGetAttribute, stream, attr, value_out); if (value_out == nullptr) { - HIP_RETURN(hipErrorInvalidValue); + return hipErrorInvalidValue; } if (!hip::isValid(stream)) {