From 2eb8cc8e90e6e453c2de3e98d2f599b78cc3d76b Mon Sep 17 00:00:00 2001 From: kjayapra-amd Date: Mon, 13 Apr 2020 22:51:46 -0400 Subject: [PATCH] SWDEV-229480 - Improve error messages in HIP Layer. Change-Id: I054b979d3aa6cf6ed4ca14a9393bdcba757772ff --- vdi/hip_context.cpp | 1 + vdi/hip_memory.cpp | 42 ++++++++++++++++++-- vdi/hip_module.cpp | 11 ++++++ vdi/hip_peer.cpp | 0 vdi/hip_platform.cpp | 39 +++++++++++++++++- vdi/hip_rtc.cpp | 4 ++ vdi/hip_texture.cpp | 94 ++++++++++++++++++-------------------------- 7 files changed, 131 insertions(+), 60 deletions(-) mode change 100644 => 100755 vdi/hip_context.cpp mode change 100644 => 100755 vdi/hip_memory.cpp mode change 100644 => 100755 vdi/hip_peer.cpp mode change 100644 => 100755 vdi/hip_rtc.cpp mode change 100644 => 100755 vdi/hip_texture.cpp diff --git a/vdi/hip_context.cpp b/vdi/hip_context.cpp old mode 100644 new mode 100755 index 440c3f4b47..1e2ae46fd5 --- a/vdi/hip_context.cpp +++ b/vdi/hip_context.cpp @@ -216,6 +216,7 @@ hipError_t hipCtxPopCurrent(hipCtx_t* ctx) { *dev = g_ctxtStack.top(); g_ctxtStack.pop(); } else { + DevLogError("Context Stack empty \n"); HIP_RETURN(hipErrorInvalidContext); } diff --git a/vdi/hip_memory.cpp b/vdi/hip_memory.cpp old mode 100644 new mode 100755 index eb56b69bd2..9debd91bf6 --- a/vdi/hip_memory.cpp +++ b/vdi/hip_memory.cpp @@ -179,8 +179,7 @@ hipError_t ihipMemcpy(void* dst, const void* src, size_t sizeBytes, hipMemcpyKin hipError_t hipExtMallocWithFlags(void** ptr, size_t sizeBytes, unsigned int flags) { HIP_INIT_API(hipExtMallocWithFlags, ptr, sizeBytes, flags); - if (flags != hipDeviceMallocDefault && - flags != hipDeviceMallocFinegrained) { + if (flags != hipDeviceMallocDefault && flags != hipDeviceMallocFinegrained) { HIP_RETURN(hipErrorInvalidValue); } @@ -205,6 +204,9 @@ hipError_t hipHostMalloc(void** ptr, size_t sizeBytes, unsigned int flags) { // can't have both Coherent and NonCoherent flags set at the same time if ((flags & coherentFlags) == coherentFlags) { + DevLogPrintfError("Cannot have both coherent and non-coherent flags " + "at the same time, flags: %u coherent flags: %u \n", + flags, coherentFlags); HIP_RETURN(hipErrorInvalidValue); } @@ -417,16 +419,20 @@ amd::Image* ihipImageCreate(const cl_channel_order channelOrder, amd::Memory* buffer) { const amd::Image::Format imageFormat({channelOrder, channelType}); if (!imageFormat.isValid()) { + DevLogPrintfError("Invalid Image format for channel Order:%u Type:%u \n", + channelOrder, channelType); return nullptr; } amd::Context& context = *hip::getCurrentDevice()->asContext(); if (!imageFormat.isSupported(context, imageType)) { + DevLogPrintfError("Image type: %u not supported \n", imageType); return nullptr; } const std::vector& devices = context.devices(); if (!devices[0]->info().imageSupport_) { + DevLogPrintfError("Device: 0x%x does not support image \n", devices[0]); return nullptr; } @@ -436,6 +442,7 @@ amd::Image* ihipImageCreate(const cl_channel_order channelOrder, imageHeight, imageDepth, imageArraySize)) { + DevLogError("Image does not have valid dimensions \n"); return nullptr; } @@ -509,6 +516,7 @@ amd::Image* ihipImageCreate(const cl_channel_order channelOrder, } if (!image->create(nullptr)) { + DevLogPrintfError("Cannot create image: 0x%x \n", image); delete image; return nullptr; } @@ -630,8 +638,7 @@ hipError_t hipMalloc3DArray(hipArray_t* array, hipError_t hipHostGetFlags(unsigned int* flagsPtr, void* hostPtr) { HIP_INIT_API(hipHostGetFlags, flagsPtr, hostPtr); - if (flagsPtr == nullptr || - hostPtr == nullptr) { + if (flagsPtr == nullptr || hostPtr == nullptr) { HIP_RETURN(hipErrorInvalidValue); } @@ -657,6 +664,8 @@ hipError_t hipHostRegister(void* hostPtr, size_t sizeBytes, unsigned int flags) constexpr bool forceAlloc = true; if (!mem->create(hostPtr, sysMemAlloc, skipAlloc, forceAlloc)) { mem->release(); + DevLogPrintfError("Cannot create memory for size: %u with flags: %d \n", + sizeBytes, flags); HIP_RETURN(hipErrorOutOfMemory); } @@ -703,6 +712,7 @@ hipError_t hipHostUnregister(void* hostPtr) { } } + DevLogPrintfError("Cannot unregister host_ptr: 0x%x \n", hostPtr); HIP_RETURN(hipErrorInvalidValue); } @@ -721,16 +731,20 @@ hipError_t hipMemcpyToSymbol(const void* symbol, const void* src, size_t count, std::string symbolName; if (!PlatformState::instance().findSymbol(symbol, symbolName)) { + DevLogPrintfError("cannot find symbol 0x%x \n", symbolName.c_str()); HIP_RETURN(hipErrorInvalidSymbol); } /* Get address and size for the global symbol */ if (!PlatformState::instance().getGlobalVar(symbolName.c_str(), ihipGetDevice(), nullptr, &device_ptr, &sym_size)) { + DevLogPrintfError("Cannot get global var: %s at device: %d \n", symbolName.c_str(), ihipGetDevice()); HIP_RETURN(hipErrorInvalidSymbol); } /* Size Check to make sure offset is correct */ if ((offset + count) != sym_size) { + DevLogPrintfError("Size does not match, offset: %u count: %u sym_size: %u \n", + offset, count, sym_size); return HIP_RETURN(hipErrorInvalidDevicePointer); } @@ -749,16 +763,20 @@ hipError_t hipMemcpyFromSymbol(void* dst, const void* symbol, size_t count, std::string symbolName; if (!PlatformState::instance().findSymbol(symbol, symbolName)) { + DevLogPrintfError("cannot find symbol: 0x%x \n", symbol); HIP_RETURN(hipErrorInvalidSymbol); } /* Get address and size for the global symbol */ if (!PlatformState::instance().getGlobalVar(symbolName.c_str(), ihipGetDevice(), nullptr, &device_ptr, &sym_size)) { + DevLogPrintfError("Cannot find symbol Name: %s \n", symbolName.c_str()); HIP_RETURN(hipErrorInvalidSymbol); } /* Size Check to make sure offset is correct */ if ((offset + count) != sym_size) { + DevLogPrintfError("Size does not match, offset: %u count: %u sym_size: %u \n", + offset, count, sym_size); return HIP_RETURN(hipErrorInvalidDevicePointer); } @@ -777,16 +795,20 @@ hipError_t hipMemcpyToSymbolAsync(const void* symbol, const void* src, size_t co std::string symbolName; if (!PlatformState::instance().findSymbol(symbol, symbolName)) { + DevLogPrintfError("cannot find symbol: 0x%x \n", symbol); HIP_RETURN(hipErrorInvalidSymbol); } /* Get address and size for the global symbol */ if (!PlatformState::instance().getGlobalVar(symbolName.c_str(), ihipGetDevice(), nullptr, &device_ptr, &sym_size)) { + DevLogPrintfError("Cannot find symbol Name: %s \n", symbolName.c_str()); HIP_RETURN(hipErrorInvalidSymbol); } /* Size Check to make sure offset is correct */ if ((offset + count) != sym_size) { + DevLogPrintfError("Size does not match, offset: %u count: %u sym_size: %u \n", + offset, count, sym_size); return HIP_RETURN(hipErrorInvalidDevicePointer); } @@ -805,16 +827,20 @@ hipError_t hipMemcpyFromSymbolAsync(void* dst, const void* symbol, size_t count, std::string symbolName; if (!PlatformState::instance().findSymbol(symbol, symbolName)) { + DevLogPrintfError("cannot find symbol: 0x%x \n", symbol); HIP_RETURN(hipErrorInvalidSymbol); } /* Get address and size for the global symbol */ if (!PlatformState::instance().getGlobalVar(symbolName.c_str(), ihipGetDevice(), nullptr, &device_ptr, &sym_size)) { + DevLogPrintfError("Cannot find symbol Name: %s \n", symbolName.c_str()); HIP_RETURN(hipErrorInvalidSymbol); } /* Size Check to make sure offset is correct */ if ((offset + count) != sym_size) { + DevLogPrintfError("Size does not match, offset: %u count: %u sym_size: %u \n", + offset, count, sym_size); return HIP_RETURN(hipErrorInvalidDevicePointer); } @@ -1837,12 +1863,16 @@ hipError_t hipIpcGetMemHandle(hipIpcMemHandle_t* handle, void* dev_ptr) { /* Get AMD::Memory object corresponding to this pointer */ amd_mem_obj = getMemoryObject(dev_ptr, offset); if (amd_mem_obj == nullptr) { + DevLogPrintfError("Cannot retrieve amd_mem_obj for dev_ptr: 0x%x with offset: %u \n", + dev_ptr, offset); HIP_RETURN(hipErrorInvalidDevicePointer); } /* Get Device::Memory object pointer */ dev_mem_obj = amd_mem_obj->getDeviceMemory(*hip::getCurrentDevice()->devices()[0],false); if (dev_mem_obj == nullptr) { + DevLogPrintfError("Cannot get Device memory for amd_mem_obj: 0x%x dev_ptr: 0x%x offset: %u \n", + amd_mem_obj, dev_ptr, offset); HIP_RETURN(hipErrorInvalidDevicePointer); } @@ -1870,6 +1900,7 @@ hipError_t hipIpcOpenMemHandle(void** dev_ptr, hipIpcMemHandle_t handle, unsigne amd_mem_obj = device->IpcAttach(&(ihandle->ipc_handle), ihandle->psize, flags, dev_ptr); if (amd_mem_obj == nullptr) { + DevLogPrintfError("cannot attach ipc_handle: with ipc_size: %u flags: %u", ihandle->psize, flags); HIP_RETURN(hipErrorInvalidDevicePointer); } @@ -1954,9 +1985,12 @@ hipError_t hipPointerGetAttributes(hipPointerAttribute_t* attributes, const void } ++device; } + DevLogPrintfError("Cannot find memory object context, memObjCtx: 0x%x \n", + memObjCtx); HIP_RETURN(hipErrorInvalidDevice); } + DevLogPrintfError("Cannot get amd_mem_obj for ptr: 0x%x \n", ptr); HIP_RETURN(hipErrorInvalidValue); } diff --git a/vdi/hip_module.cpp b/vdi/hip_module.cpp index 3d40d8c967..f523a86a0f 100755 --- a/vdi/hip_module.cpp +++ b/vdi/hip_module.cpp @@ -94,10 +94,12 @@ hipError_t hipModuleUnload(hipModule_t hmod) amd::Program* program = as_amd(reinterpret_cast(hmod)); if(!PlatformState::instance().unregisterFunc(hmod)) { + DevLogPrintfError("Cannot unregister module: 0x%x \n", hmod); HIP_RETURN(hipErrorInvalidSymbol); } if(!ihipModuleUnregisterGlobal(hmod)) { + DevLogPrintfError("Cannot unregister Global vars for module: 0x%x \n", hmod); HIP_RETURN(hipErrorInvalidSymbol); } @@ -134,6 +136,7 @@ inline bool ihipModuleRegisterUndefined(amd::Program* program, hipModule_t* modu = program->getDeviceProgram(*hip::getCurrentDevice()->devices()[0]); if (!dev_program->getUndefinedVarFromCodeObj(&undef_vars)) { + DevLogPrintfError("Could not get undefined Variables for Module: 0x%x \n", *module); return false; } @@ -163,6 +166,7 @@ inline bool ihipModuleRegisterFunc(amd::Program* program, hipModule_t* module) { // Get all the global func names from COMGR if (!dev_program->getGlobalFuncFromCodeObj(&func_names)) { + DevLogPrintfError("Could not get Global Funcs from Code Obj for Module: 0x%x \n", *module); return false; } @@ -180,6 +184,7 @@ inline bool ihipModuleRegisterGlobal(amd::Program* program, hipModule_t* module) = program->getDeviceProgram(*hip::getCurrentDevice()->devices()[0]); if (!dev_program->getGlobalVarFromCodeObj(&var_names)) { + DevLogPrintfError("Could not get Global vars from Code Obj for Module: 0x%x \n", *module); return false; } @@ -244,6 +249,8 @@ hipError_t hipModuleGetFunction(hipFunction_t *hfunc, hipModule_t hmod, const ch HIP_INIT_API(hipModuleGetFunction, hfunc, hmod, name); if (!PlatformState::instance().findModFunc(hfunc, hmod, name)) { + DevLogPrintfError("Cannot find the function: %s for module: 0x%x \n", + name, hmod); HIP_RETURN(hipErrorNotFound); } HIP_RETURN(hipSuccess); @@ -256,6 +263,8 @@ hipError_t hipModuleGetGlobal(hipDeviceptr_t* dptr, size_t* bytes, hipModule_t h /* Get address and size for the global symbol */ if (!PlatformState::instance().getGlobalVar(name, ihipGetDevice(), hmod, dptr, bytes)) { + DevLogPrintfError("Cannot find global Var: %s for module: 0x%x at device: %d \n", + name, hmod, ihipGetDevice()); HIP_RETURN(hipErrorNotFound); } @@ -653,6 +662,8 @@ hipError_t hipModuleGetTexRef(textureReference** texRef, hipModule_t hmod, const /* Get address and size for the global symbol */ if (!PlatformState::instance().getTexRef(name, hmod, texRef)) { + DevLogPrintfError("Cannot get texRef for name: %s at module:0x%x \n", + name, hmod); HIP_RETURN(hipErrorNotFound); } diff --git a/vdi/hip_peer.cpp b/vdi/hip_peer.cpp old mode 100644 new mode 100755 diff --git a/vdi/hip_platform.cpp b/vdi/hip_platform.cpp index 822f68dce6..7b63d5225b 100755 --- a/vdi/hip_platform.cpp +++ b/vdi/hip_platform.cpp @@ -132,6 +132,8 @@ extern "C" std::vector>* __hipRegisterFatBinary(con { const __CudaFatBinaryWrapper* fbwrapper = reinterpret_cast(data); if (fbwrapper->magic != __hipFatMAGIC2 || fbwrapper->version != 1) { + DevLogPrintfError("Cannot Register fat binary. FatMagic: %u version: %u ", + fbwrapper->magic, fbwrapper->version); return nullptr; } @@ -272,6 +274,7 @@ bool PlatformState::findSymbol(const void *hostVar, std::string &symbolName) { symbolName = it->second; return true; } + DevLogPrintfError("Could not find the Symbol: %s \n", symbolName.c_str()); return false; } @@ -298,6 +301,7 @@ bool ihipGetFuncAttributes(const char* func_name, amd::Program* program, hipFunc const auto it = dev_program->kernels().find(std::string(func_name)); if (it == dev_program->kernels().cend()) { + DevLogPrintfError("Could not find the function %s \n", func_name); return false; } @@ -318,6 +322,7 @@ bool PlatformState::getShadowVarInfo(std::string var_name, hipModule_t hmod, *var_size = dvar->size; return true; } else { + DevLogPrintfError("Cannot find Var name: %s in module: 0x%x \n", var_name.c_str(), hmod); return false; } } @@ -357,6 +362,7 @@ bool PlatformState::findModFunc(hipFunction_t* hfunc, hipModule_t hmod, const ch PlatformState::DeviceFunction& devFunc = func_it->second; if (devFunc.functions[ihipGetDevice()] == 0) { if(!createFunc(&devFunc.functions[ihipGetDevice()], hmod, name)) { + DevLogPrintfError("Could not create a function: %s at module: 0x%x \n", name, hmod); return false; } } @@ -364,6 +370,7 @@ bool PlatformState::findModFunc(hipFunction_t* hfunc, hipModule_t hmod, const ch return true; } } + DevLogPrintfError("Cannot find module: 0x%x in PlatformState Module Map \n", hmod); return false; } @@ -372,15 +379,22 @@ bool PlatformState::createFunc(hipFunction_t* hfunc, hipModule_t hmod, const cha const amd::Symbol* symbol = program->findSymbol(name); if (!symbol) { + DevLogPrintfError("Cannot find Symbol with name: %s \n", name); return false; } amd::Kernel* kernel = new amd::Kernel(*program, *symbol, name); if (!kernel) { + DevLogPrintfError("Could not create a new kernel with name: %s \n", name); return false; } hip::Function* f = new hip::Function(kernel); + if (!f) { + DevLogPrintfError("Could not create a new function with name: %s \n", name); + return false; + } + *hfunc = f->asHipFunction(); return true; @@ -398,6 +412,7 @@ hipFunction_t PlatformState::getFunc(const void* hostFunction, int deviceId) { amd::Program* program = as_amd(reinterpret_cast(module)); program->setVarInfoCallBack(&getSvarInfo); if (CL_SUCCESS != program->build(g_devices[deviceId]->devices(), nullptr, nullptr, nullptr)) { + DevLogPrintfError("Build error for module: 0x%x at device: %u \n", module, deviceId); return nullptr; } (*devFunc.modules)[deviceId].second = true; @@ -414,6 +429,7 @@ hipFunction_t PlatformState::getFunc(const void* hostFunction, int deviceId) { } return devFunc.functions[deviceId]; } + DevLogPrintfError("Cannot find function: 0x%x in PlatformState \n", hostFunction); return nullptr; } @@ -425,6 +441,7 @@ bool PlatformState::getFuncAttr(const void* hostFunction, const auto it = functions_.find(hostFunction); if (it == functions_.cend()) { + DevLogPrintfError("Cannot find hostFunction 0x%x \n", hostFunction); return false; } @@ -434,12 +451,15 @@ bool PlatformState::getFuncAttr(const void* hostFunction, /* If module has not been initialized yet, build the kernel now*/ if (!(*devFunc.modules)[deviceId].second) { if (nullptr == PlatformState::instance().getFunc(hostFunction, deviceId)) { + DevLogPrintfError("Cannot get hostFunction: 0x%x for deviceId:%d \n", hostFunction, deviceId); return false; } } amd::Program* program = as_amd(reinterpret_cast((*devFunc.modules)[deviceId].first)); if (!ihipGetFuncAttributes(devFunc.deviceName.c_str(), program, func_attr)) { + DevLogPrintfError("Cannot get Func attributes for function: %s \n", + devFunc.deviceName.c_str()); return false; } return true; @@ -449,10 +469,13 @@ bool PlatformState::getTexRef(const char* hostVar, hipModule_t hmod, textureRefe amd::ScopedLock lock(lock_); DeviceVar* dvar = findVar(std::string(hostVar), ihipGetDevice(), hmod); if (dvar == nullptr) { + DevLogPrintfError("Cannot find var:%s for creating texture reference at module: 0x%x \n", + hostVar, hmod); return false; } if (!dvar->dyn_undef) { + DevLogPrintfError("HostVar: %s is not created through hipModuleLoad \n", hostVar); return false; } @@ -475,6 +498,7 @@ bool PlatformState::getGlobalVar(const char* hostVar, int deviceId, hipModule_t amd::Program* program = as_amd(reinterpret_cast((*dvar->modules)[deviceId].first)); program->setVarInfoCallBack(&getSvarInfo); if (CL_SUCCESS != program->build(g_devices[deviceId]->devices(), nullptr, nullptr, nullptr)) { + DevLogPrintfError("Build Failure for module: 0x%x \n", hmod); return false; } (*dvar->modules)[deviceId].second = true; @@ -487,13 +511,14 @@ bool PlatformState::getGlobalVar(const char* hostVar, int deviceId, hipModule_t dvar->rvars[deviceId].amd_mem_obj_ = amd_mem_obj; amd::MemObjMap::AddMemObj(device_ptr, amd_mem_obj); } else { - LogError("[HIP] __hipRegisterVar cannot find kernel for device \n"); + LogError("__hipRegisterVar cannot find kernel for device \n"); } } *size_ptr = dvar->rvars[deviceId].getvarsize(); *dev_ptr = dvar->rvars[deviceId].getdeviceptr(); return true; } else { + DevLogPrintfError("Could not find global var: %s at module:0x%x \n", hostVar, hmod); return false; } } @@ -638,10 +663,12 @@ extern "C" hipError_t hipLaunchByPtr(const void *hostFunction) hip::Stream* stream = reinterpret_cast(exec.hStream_); int deviceId = (stream != nullptr)? stream->device->deviceId() : ihipGetDevice(); if (deviceId == -1) { + DevLogPrintfError("Wrong DeviceId: %d \n", deviceId); HIP_RETURN(hipErrorNoDevice); } hipFunction_t func = PlatformState::instance().getFunc(hostFunction, deviceId); if (func == nullptr) { + DevLogPrintfError("Could not retrieve hostFunction: 0x%x \n", hostFunction); HIP_RETURN(hipErrorInvalidDeviceFunction); } @@ -663,11 +690,14 @@ hipError_t hipGetSymbolAddress(void** devPtr, const void* symbol) { std::string symbolName; if (!PlatformState::instance().findSymbol(symbol, symbolName)) { + DevLogPrintfError("Cannot find symbol: %s \n", symbolName.c_str()); HIP_RETURN(hipErrorInvalidSymbol); } size_t size = 0; if(!PlatformState::instance().getGlobalVar(symbolName.c_str(), ihipGetDevice(), nullptr, devPtr, &size)) { + DevLogPrintfError("Cannot find global variable device ptr for symbol: %s at device: %d \n", + symbolName.c_str(), ihipGetDevice()); HIP_RETURN(hipErrorInvalidSymbol); } HIP_RETURN(hipSuccess); @@ -678,11 +708,14 @@ hipError_t hipGetSymbolSize(size_t* sizePtr, const void* symbol) { std::string symbolName; if (!PlatformState::instance().findSymbol(symbol, symbolName)) { + DevLogPrintfError("Cannot find symbol: %s \n", symbolName.c_str()); HIP_RETURN(hipErrorInvalidSymbol); } hipDeviceptr_t devPtr = nullptr; if (!PlatformState::instance().getGlobalVar(symbolName.c_str(), ihipGetDevice(), nullptr, &devPtr, sizePtr)) { + DevLogPrintfError("Cannot find global variable device ptr for symbol: %s at device: %d \n", + symbolName.c_str(), ihipGetDevice()); HIP_RETURN(hipErrorInvalidSymbol); } HIP_RETURN(hipSuccess); @@ -701,10 +734,12 @@ hipError_t ihipCreateGlobalVarObj(const char* name, hipModule_t hmod, amd::Memor dev_program = program->getDeviceProgram(*hip::getCurrentDevice()->devices()[0]); if (dev_program == nullptr) { + DevLogPrintfError("Cannot get Device Function for module: 0x%x \n", hmod); HIP_RETURN(hipErrorInvalidDeviceFunction); } /* Find the global Symbols */ if (!dev_program->createGlobalVarObj(amd_mem_obj, dptr, bytes, name)) { + DevLogPrintfError("Cannot create Global Var obj for symbol: %s \n", name); HIP_RETURN(hipErrorInvalidSymbol); } @@ -1107,6 +1142,7 @@ extern "C" hipError_t hipLaunchKernel(const void *hostFunction, hip::Stream* s = reinterpret_cast(stream); int deviceId = (s != nullptr)? s->device->deviceId() : ihipGetDevice(); if (deviceId == -1) { + DevLogPrintfError("Wrong Device Id: %d \n", deviceId); HIP_RETURN(hipErrorNoDevice); } hipFunction_t func = PlatformState::instance().getFunc(hostFunction, deviceId); @@ -1114,6 +1150,7 @@ extern "C" hipError_t hipLaunchKernel(const void *hostFunction, #ifdef ATI_OS_LINUX const auto it = hip_impl::functions().find(reinterpret_cast(hostFunction)); if (it == hip_impl::functions().cend()) { + DevLogPrintfError("Cannot find function: 0x%x \n", hostFunction); HIP_RETURN(hipErrorInvalidDeviceFunction); } func = it->second; diff --git a/vdi/hip_rtc.cpp b/vdi/hip_rtc.cpp old mode 100644 new mode 100755 index 9897b98b7f..8c82337405 --- a/vdi/hip_rtc.cpp +++ b/vdi/hip_rtc.cpp @@ -109,6 +109,7 @@ char* demangle(const char* loweredName) { int status = 0; char* demangledName = DEMANGLE(loweredName, nullptr, nullptr, &status); if (status != 0) { + DevLogPrintfError("Cannot demangle loweredName: %s \n", loweredName); return nullptr; } #elif defined(_WIN32) @@ -118,6 +119,8 @@ char* demangle(const char* loweredName) { UNDECORATED_SIZE/ sizeof(*demangledName), UNDNAME_COMPLETE)) { free(demangledName); + DevLogPrintfError("Cannot undecorate loweredName: %s demangledName: %s \n", + loweredName, demangedName); return nullptr; } #else @@ -192,6 +195,7 @@ const char* hiprtcGetErrorString(hiprtcResult x) { case HIPRTC_ERROR_INTERNAL_ERROR: return "HIPRTC_ERROR_INTERNAL_ERROR"; default: + DevLogPrintfError("Invalid HIPRTC error code: %d \n", x); return nullptr; }; diff --git a/vdi/hip_texture.cpp b/vdi/hip_texture.cpp old mode 100644 new mode 100755 index b837729721..da24d663d1 --- a/vdi/hip_texture.cpp +++ b/vdi/hip_texture.cpp @@ -335,8 +335,7 @@ hipError_t hipGetTextureObjectResourceDesc(hipResourceDesc* pResDesc, hipTextureObject_t texObject) { HIP_INIT_API(hipGetTextureObjectResourceDesc, pResDesc, texObject); - if ((pResDesc == nullptr) || - (texObject == nullptr)) { + if ((pResDesc == nullptr) || (texObject == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -349,8 +348,7 @@ hipError_t hipGetTextureObjectResourceViewDesc(hipResourceViewDesc* pResViewDesc hipTextureObject_t texObject) { HIP_INIT_API(hipGetTextureObjectResourceViewDesc, pResViewDesc, texObject); - if ((pResViewDesc == nullptr) || - (texObject == nullptr)) { + if ((pResViewDesc == nullptr) || (texObject == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -363,8 +361,7 @@ hipError_t hipGetTextureObjectTextureDesc(hipTextureDesc* pTexDesc, hipTextureObject_t texObject) { HIP_INIT_API(hipGetTextureObjectTextureDesc, pTexDesc, texObject); - if ((pTexDesc == nullptr) || - (texObject == nullptr)) { + if ((pTexDesc == nullptr) || (texObject == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -383,8 +380,8 @@ inline bool ihipGetTextureAlignmentOffset(size_t* offset, // If the device memory pointer was returned from hipMalloc(), // the offset is guaranteed to be 0 and NULL may be passed as the offset parameter. - if ((alignedOffset != 0) && - (offset == nullptr)) { + if ((alignedOffset != 0) && (offset == nullptr)) { + DevLogPrintfError("Texture object not aligned with offset %u \n", alignedOffset); return false; } @@ -574,8 +571,7 @@ hipError_t hipGetChannelDesc(hipChannelFormatDesc* desc, hipArray_const_t array) { HIP_INIT_API(hipGetChannelDesc, desc, array); - if ((desc == nullptr) || - (array == nullptr)) { + if ((desc == nullptr) || (array == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -590,8 +586,7 @@ hipError_t hipGetTextureAlignmentOffset(size_t* offset, const textureReference* texref) { HIP_INIT_API(hipGetTextureAlignmentOffset, offset, texref); - if ((offset == nullptr) || - (texref == nullptr)) { + if ((offset == nullptr) || (texref == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -670,13 +665,14 @@ hipError_t hipTexRefGetAddressMode(hipTextureAddressMode* pam, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetAddressMode, pam, texRef, dim); - if ((pam == nullptr) || - (texRef == nullptr)) { + if ((pam == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } // Currently, the only valid value for dim are 0 and 1. if ((dim != 0) || (dim != 1)) { + DevLogPrintfError("Currently only 2 dimensions (0,1) are valid," + "dim : %d \n", dim); HIP_RETURN(hipErrorInvalidValue); } @@ -695,6 +691,8 @@ hipError_t hipTexRefSetAddressMode(textureReference* texRef, } if ((dim < 0) || (dim > 2)) { + DevLogPrintfError("Currently only 3 dimensions (0,1,2) are valid," + "dim : %d \n", dim); HIP_RETURN(hipErrorInvalidValue); } @@ -708,8 +706,7 @@ hipError_t hipTexRefGetArray(hipArray_t* pArray, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetArray, pArray, texRef); - if ((pArray == nullptr) || - (texRef == nullptr)) { + if ((pArray == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -723,8 +720,9 @@ hipError_t hipTexRefGetArray(hipArray_t* pArray, switch (resDesc.resType) { case hipResourceTypeLinear: case hipResourceTypePitch2D: - case hipResourceTypeMipmappedArray: + case hipResourceTypeMipmappedArray: { HIP_RETURN(hipErrorInvalidValue); + } case hipResourceTypeArray: *pArray = resDesc.res.array.array; break; @@ -738,8 +736,7 @@ hipError_t hipTexRefSetArray(textureReference* texRef, unsigned int flags) { HIP_INIT_API(hipTexRefSetArray, texRef, array, flags); - if ((texRef == nullptr) || - (array == nullptr)) { + if ((texRef == nullptr) || (array == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -769,8 +766,7 @@ hipError_t hipTexRefGetAddress(hipDeviceptr_t* dptr, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetAddress, dptr, texRef); - if ((dptr == nullptr) || - (texRef == nullptr)) { + if ((dptr == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -778,6 +774,8 @@ hipError_t hipTexRefGetAddress(hipDeviceptr_t* dptr, // TODO use ihipGetTextureObjectResourceDesc() to not pollute the API trace. hipError_t error = hipGetTextureObjectResourceDesc(&resDesc, texRef->textureObject); if (error != hipSuccess) { + DevLogPrintfError("hipGetTextureObjectResourceDesc failed with error code: %s \n", + hipGetErrorName(error)); return HIP_RETURN(error); } @@ -786,8 +784,9 @@ hipError_t hipTexRefGetAddress(hipDeviceptr_t* dptr, // If the texture reference is not bound to any device memory range, // return hipErroInvalidValue. case hipResourceTypeArray: - case hipResourceTypeMipmappedArray: + case hipResourceTypeMipmappedArray: { HIP_RETURN(hipErrorInvalidValue); + } case hipResourceTypeLinear: *dptr = resDesc.res.linear.devPtr; break; @@ -838,8 +837,7 @@ hipError_t hipTexRefSetAddress2D(textureReference* texRef, size_t Pitch) { HIP_INIT_API(hipTexRefSetAddress2D, texRef, desc, dptr, Pitch); - if ((texRef == nullptr) || - (desc == nullptr)) { + if ((texRef == nullptr) || (desc == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -870,8 +868,7 @@ hipError_t hipTexRefGetBorderColor(float* pBorderColor, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetBorderColor, pBorderColor, texRef); - if ((pBorderColor == nullptr) || - (texRef == nullptr)) { + if ((pBorderColor == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -887,8 +884,7 @@ hipError_t hipTexRefGetFilterMode(hipTextureFilterMode* pfm, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetFilterMode, pfm, texRef); - if ((pfm == nullptr) || - (texRef == nullptr)) { + if ((pfm == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -902,8 +898,7 @@ hipError_t hipTexRefGetFlags(unsigned int* pFlags, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetFlags, pFlags, texRef); - if ((pFlags == nullptr) || - (texRef == nullptr)) { + if ((pFlags == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -930,8 +925,7 @@ hipError_t hipTexRefGetFormat(hipArray_Format* pFormat, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetFormat, pFormat, pNumChannels, texRef); - if ((pFormat == nullptr) || - (pNumChannels == nullptr) || + if ((pFormat == nullptr) || (pNumChannels == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -947,8 +941,7 @@ hipError_t hipTexRefGetMaxAnisotropy(int* pmaxAnsio, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetMaxAnisotropy, pmaxAnsio, texRef); - if ((pmaxAnsio == nullptr) || - (texRef == nullptr)) { + if ((pmaxAnsio == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -962,8 +955,7 @@ hipError_t hipTexRefGetMipmapFilterMode(hipTextureFilterMode* pfm, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetMipmapFilterMode, pfm, texRef); - if ((pfm == nullptr) || - (texRef == nullptr)) { + if ((pfm == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -977,8 +969,7 @@ hipError_t hipTexRefGetMipmapLevelBias(float* pbias, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetMipmapLevelBias, pbias, texRef); - if ((pbias == nullptr) || - (texRef == nullptr)) { + if ((pbias == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -993,8 +984,7 @@ hipError_t hipTexRefGetMipmapLevelClamp(float* pminMipmapLevelClamp, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetMipmapLevelClamp, pminMipmapLevelClamp, pmaxMipmapLevelClamp, texRef); - if ((pminMipmapLevelClamp == nullptr) || - (pmaxMipmapLevelClamp == nullptr) || + if ((pminMipmapLevelClamp == nullptr) || (pmaxMipmapLevelClamp == nullptr) || (texRef == nullptr)){ HIP_RETURN(hipErrorInvalidValue); } @@ -1010,8 +1000,7 @@ hipError_t hipTexRefGetMipmappedArray(hipMipmappedArray_t* pArray, // TODO overload operator<<(ostream&, textureReference&). HIP_INIT_API(hipTexRefGetMipmappedArray, pArray, &texRef); - if ((pArray == nullptr) || - (texRef == nullptr)) { + if ((pArray == nullptr) || (texRef == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -1025,8 +1014,9 @@ hipError_t hipTexRefGetMipmappedArray(hipMipmappedArray_t* pArray, switch (resDesc.resType) { case hipResourceTypeLinear: case hipResourceTypePitch2D: - case hipResourceTypeArray: + case hipResourceTypeArray: { HIP_RETURN(hipErrorInvalidValue); + } case hipResourceTypeMipmappedArray: *pArray = resDesc.res.mipmap.mipmap; break; @@ -1039,8 +1029,7 @@ hipError_t hipTexRefSetBorderColor(textureReference* texRef, float* pBorderColor) { HIP_INIT_API(hipTexRefSetBorderColor, texRef, pBorderColor); - if ((texRef == nullptr) || - (pBorderColor == nullptr)) { + if ((texRef == nullptr) || (pBorderColor == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -1110,8 +1099,7 @@ hipError_t hipTexRefSetMipmappedArray(textureReference* texRef, unsigned int Flags) { HIP_INIT_API(hipTexRefSetMipmappedArray, texRef, mipmappedArray, Flags); - if ((texRef == nullptr) || - (mipmappedArray == nullptr)) { + if ((texRef == nullptr) || (mipmappedArray == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -1142,8 +1130,7 @@ hipError_t hipTexObjectCreate(hipTextureObject_t* pTexObject, const HIP_RESOURCE_VIEW_DESC* pResViewDesc) { HIP_INIT_API(hipTexObjectCreate, pTexObject, pResDesc, pTexDesc, pResViewDesc); - if ((pTexObject == nullptr) || - (pResDesc == nullptr) || (pTexDesc == nullptr)) { + if ((pTexObject == nullptr) || (pResDesc == nullptr) || (pTexDesc == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -1168,8 +1155,7 @@ hipError_t hipTexObjectGetResourceDesc(HIP_RESOURCE_DESC* pResDesc, hipTextureObject_t texObject) { HIP_INIT_API(hipTexObjectGetResourceDesc, pResDesc, texObject); - if ((pResDesc == nullptr) || - (texObject == nullptr)) { + if ((pResDesc == nullptr) || (texObject == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -1182,8 +1168,7 @@ hipError_t hipTexObjectGetResourceViewDesc(HIP_RESOURCE_VIEW_DESC* pResViewDesc, hipTextureObject_t texObject) { HIP_INIT_API(hipTexObjectGetResourceViewDesc, pResViewDesc, texObject); - if ((pResViewDesc == nullptr) || - (texObject == nullptr)) { + if ((pResViewDesc == nullptr) || (texObject == nullptr)) { HIP_RETURN(hipErrorInvalidValue); } @@ -1196,8 +1181,7 @@ hipError_t hipTexObjectGetTextureDesc(HIP_TEXTURE_DESC* pTexDesc, hipTextureObject_t texObject) { HIP_INIT_API(hipTexObjectGetTextureDesc, pTexDesc, texObject); - if ((pTexDesc == nullptr) || - (texObject == nullptr)) { + if ((pTexDesc == nullptr) || (texObject == nullptr)) { HIP_RETURN(hipErrorInvalidValue); }