From e7fa085afebc21fcb91a489af48386262e193a9b Mon Sep 17 00:00:00 2001 From: Anusha GodavarthySurya Date: Wed, 30 Nov 2022 05:15:01 +0000 Subject: [PATCH] SWDEV-325711 - Correct formatting Change-Id: Ie26159e0bb3315cf7c3de1eb682f23ef343df0f2 [ROCm/clr commit: 16b31b0c5495e4cc81fce5a33185a65e63d68052] --- projects/clr/hipamd/src/hip_graph.cpp | 68 +++++++------- .../clr/hipamd/src/hip_graph_internal.cpp | 23 +++-- .../clr/hipamd/src/hip_graph_internal.hpp | 90 +++++++++---------- 3 files changed, 92 insertions(+), 89 deletions(-) diff --git a/projects/clr/hipamd/src/hip_graph.cpp b/projects/clr/hipamd/src/hip_graph.cpp index 55a1e7914e..34df77cd7d 100644 --- a/projects/clr/hipamd/src/hip_graph.cpp +++ b/projects/clr/hipamd/src/hip_graph.cpp @@ -853,8 +853,7 @@ hipError_t hipStreamBeginCapture_common(hipStream_t stream, hipStreamCaptureMode if (stream == nullptr) { return hipErrorStreamCaptureUnsupported; } - if (mode < hipStreamCaptureModeGlobal || - mode > hipStreamCaptureModeRelaxed) { + if (mode < hipStreamCaptureModeGlobal || mode > hipStreamCaptureModeRelaxed) { return hipErrorInvalidValue; } hip::Stream* s = reinterpret_cast(stream); @@ -909,8 +908,7 @@ hipError_t hipStreamEndCapture_common(hipStream_t stream, hipGraph_t* pGraph) { } // If mode is not hipStreamCaptureModeRelaxed, hipStreamEndCapture must be called on the stream // from the same thread - const auto& it = std::find(hip::tls.capture_streams_.begin(), - hip::tls.capture_streams_.end(), s); + const auto& it = std::find(hip::tls.capture_streams_.begin(), hip::tls.capture_streams_.end(), s); if (s->GetCaptureMode() != hipStreamCaptureModeRelaxed) { if (it == hip::tls.capture_streams_.end()) { return hipErrorStreamCaptureWrongThread; @@ -1118,7 +1116,8 @@ hipError_t ihipGraphInstantiate(hipGraphExec_t* pGraphExec, hipGraph_t graph) { std::vector levelOrder; clonedGraph->LevelOrder(levelOrder); clonedGraph->GetUserObjs(graphExeUserObj); - *pGraphExec = new hipGraphExec(levelOrder, parallelLists, nodeWaitLists, clonedNodes, graphExeUserObj); + *pGraphExec = + new hipGraphExec(levelOrder, parallelLists, nodeWaitLists, clonedNodes, graphExeUserObj); if (*pGraphExec != nullptr) { return (*pGraphExec)->Init(); } else { @@ -1167,7 +1166,7 @@ hipError_t ihipGraphLaunch(hipGraphExec_t graphExec, hipStream_t stream) { return graphExec->Run(stream); } -hipError_t hipGraphLaunch_common( hipGraphExec_t graphExec, hipStream_t stream) { +hipError_t hipGraphLaunch_common(hipGraphExec_t graphExec, hipStream_t stream) { if (graphExec == nullptr || !hip::isValid(stream) || !hipGraphExec::isGraphExecValid(graphExec)) { return hipErrorInvalidValue; } @@ -1270,7 +1269,8 @@ hipError_t hipGraphKernelNodeSetAttribute(hipGraphNode_t hNode, hipKernelNodeAtt if (hNode == nullptr || value == nullptr) { HIP_RETURN(hipErrorInvalidValue); } - if (attr != hipKernelNodeAttributeAccessPolicyWindow && attr != hipKernelNodeAttributeCooperative ) { + if (attr != hipKernelNodeAttributeAccessPolicyWindow && + attr != hipKernelNodeAttributeCooperative) { HIP_RETURN(hipErrorInvalidValue); } HIP_RETURN(reinterpret_cast(hNode)->SetAttrParams(attr, value)); @@ -1282,7 +1282,8 @@ hipError_t hipGraphKernelNodeGetAttribute(hipGraphNode_t hNode, hipKernelNodeAtt if (hNode == nullptr || value == nullptr) { HIP_RETURN(hipErrorInvalidValue); } - if (attr != hipKernelNodeAttributeAccessPolicyWindow && attr != hipKernelNodeAttributeCooperative ) { + if (attr != hipKernelNodeAttributeAccessPolicyWindow && + attr != hipKernelNodeAttributeCooperative) { HIP_RETURN(hipErrorInvalidValue); } HIP_RETURN(reinterpret_cast(hNode)->GetAttrParams(attr, value)); @@ -1423,7 +1424,7 @@ hipError_t hipGraphExecChildGraphNodeSetParams(hipGraphExec_t hGraphExec, hipGra HIP_RETURN(hipErrorUnknown); } -// Validate whether the topology of node and childGraph matches + // Validate whether the topology of node and childGraph matches std::vector childGraphNodes1; node->LevelOrder(childGraphNodes1); @@ -1449,8 +1450,9 @@ hipError_t hipGraphExecChildGraphNodeSetParams(hipGraphExec_t hGraphExec, hipGra HIP_RETURN(reinterpret_cast(clonedNode)->SetParams(childGraph)); } -hipError_t hipStreamGetCaptureInfo_common(hipStream_t stream, hipStreamCaptureStatus* pCaptureStatus, - unsigned long long* pId) { +hipError_t hipStreamGetCaptureInfo_common(hipStream_t stream, + hipStreamCaptureStatus* pCaptureStatus, + unsigned long long* pId) { if (pCaptureStatus == nullptr || !hip::isValid(stream)) { return hipErrorInvalidValue; } @@ -1472,16 +1474,17 @@ hipError_t hipStreamGetCaptureInfo(hipStream_t stream, hipStreamCaptureStatus* p } hipError_t hipStreamGetCaptureInfo_spt(hipStream_t stream, hipStreamCaptureStatus* pCaptureStatus, - unsigned long long* pId) { + unsigned long long* pId) { HIP_INIT_API(hipStreamGetCaptureInfo, stream, pCaptureStatus, pId); PER_THREAD_DEFAULT_STREAM(stream); HIP_RETURN(hipStreamGetCaptureInfo_common(stream, pCaptureStatus, pId)); } -hipError_t hipStreamGetCaptureInfo_v2_common(hipStream_t stream, hipStreamCaptureStatus* captureStatus_out, - unsigned long long* id_out, hipGraph_t* graph_out, - const hipGraphNode_t** dependencies_out, - size_t* numDependencies_out) { +hipError_t hipStreamGetCaptureInfo_v2_common(hipStream_t stream, + hipStreamCaptureStatus* captureStatus_out, + unsigned long long* id_out, hipGraph_t* graph_out, + const hipGraphNode_t** dependencies_out, + size_t* numDependencies_out) { if (captureStatus_out == nullptr) { return hipErrorInvalidValue; } @@ -1517,18 +1520,19 @@ hipError_t hipStreamGetCaptureInfo_v2(hipStream_t stream, hipStreamCaptureStatus HIP_INIT_API(hipStreamGetCaptureInfo_v2, stream, captureStatus_out, id_out, graph_out, dependencies_out, numDependencies_out); HIP_RETURN(hipStreamGetCaptureInfo_v2_common(stream, captureStatus_out, id_out, graph_out, - dependencies_out, numDependencies_out)); + dependencies_out, numDependencies_out)); } -hipError_t hipStreamGetCaptureInfo_v2_spt(hipStream_t stream, hipStreamCaptureStatus* captureStatus_out, - unsigned long long* id_out, hipGraph_t* graph_out, - const hipGraphNode_t** dependencies_out, - size_t* numDependencies_out) { +hipError_t hipStreamGetCaptureInfo_v2_spt(hipStream_t stream, + hipStreamCaptureStatus* captureStatus_out, + unsigned long long* id_out, hipGraph_t* graph_out, + const hipGraphNode_t** dependencies_out, + size_t* numDependencies_out) { HIP_INIT_API(hipStreamGetCaptureInfo_v2, stream, captureStatus_out, id_out, graph_out, dependencies_out, numDependencies_out); PER_THREAD_DEFAULT_STREAM(stream); HIP_RETURN(hipStreamGetCaptureInfo_v2_common(stream, captureStatus_out, id_out, graph_out, - dependencies_out, numDependencies_out)); + dependencies_out, numDependencies_out)); } hipError_t hipStreamUpdateCaptureDependencies(hipStream_t stream, hipGraphNode_t* dependencies, @@ -1947,8 +1951,8 @@ hipError_t hipGraphHostNodeGetParams(hipGraphNode_t node, hipHostNodeParams* pNo hipError_t hipGraphHostNodeSetParams(hipGraphNode_t node, const hipHostNodeParams* pNodeParams) { HIP_INIT_API(hipGraphHostNodeSetParams, node, pNodeParams); - if (pNodeParams == nullptr || pNodeParams->fn == nullptr || - node == nullptr || !hipGraphNode::isNodeValid(node)) { + if (pNodeParams == nullptr || pNodeParams->fn == nullptr || node == nullptr || + !hipGraphNode::isNodeValid(node)) { HIP_RETURN(hipErrorInvalidValue); } HIP_RETURN(reinterpret_cast(node)->SetParams(pNodeParams)); @@ -1957,7 +1961,8 @@ hipError_t hipGraphHostNodeSetParams(hipGraphNode_t node, const hipHostNodeParam hipError_t hipGraphExecHostNodeSetParams(hipGraphExec_t hGraphExec, hipGraphNode_t node, const hipHostNodeParams* pNodeParams) { HIP_INIT_API(hipGraphExecHostNodeSetParams, hGraphExec, node, pNodeParams); - if (hGraphExec == nullptr || pNodeParams == nullptr || pNodeParams->fn == nullptr || node == nullptr || !hipGraphNode::isNodeValid(node)) { + if (hGraphExec == nullptr || pNodeParams == nullptr || pNodeParams->fn == nullptr || + node == nullptr || !hipGraphNode::isNodeValid(node)) { HIP_RETURN(hipErrorInvalidValue); } hipGraphNode_t clonedNode = hGraphExec->GetClonedNode(node); @@ -2061,9 +2066,11 @@ hipError_t hipDeviceGraphMemTrim(int device) { return HIP_RETURN(hipSuccess); } -hipError_t hipUserObjectCreate(hipUserObject_t* object_out, void* ptr, hipHostFn_t destroy, unsigned int initialRefcount, unsigned int flags) { +hipError_t hipUserObjectCreate(hipUserObject_t* object_out, void* ptr, hipHostFn_t destroy, + unsigned int initialRefcount, unsigned int flags) { HIP_INIT_API(hipUserObjectCreate, object_out, ptr, destroy, initialRefcount, flags); - if (object_out == nullptr || flags != hipUserObjectNoDestructorSync || initialRefcount == 0 || destroy == nullptr || initialRefcount > INT_MAX) { + if (object_out == nullptr || flags != hipUserObjectNoDestructorSync || initialRefcount == 0 || + destroy == nullptr || initialRefcount > INT_MAX) { HIP_RETURN(hipErrorInvalidValue); } @@ -2099,7 +2106,8 @@ hipError_t hipUserObjectRetain(hipUserObject_t object, unsigned int count) { HIP_RETURN(hipSuccess); } -hipError_t hipGraphRetainUserObject(hipGraph_t graph, hipUserObject_t object, unsigned int count, unsigned int flags) { +hipError_t hipGraphRetainUserObject(hipGraph_t graph, hipUserObject_t object, unsigned int count, + unsigned int flags) { HIP_INIT_API(hipGraphRetainUserObject, graph, object, count, flags); hipError_t status = hipSuccess; if (graph == nullptr || object == nullptr || count == 0 || count > INT_MAX || @@ -2141,7 +2149,7 @@ hipError_t hipGraphReleaseUserObject(hipGraph_t graph, hipUserObject_t object, u hipError_t hipGraphKernelNodeCopyAttributes(hipGraphNode_t hSrc, hipGraphNode_t hDst) { HIP_INIT_API(hipGraphKernelNodeCopyAttributes, hSrc, hDst); - if(hSrc == nullptr || hDst == nullptr) { + if (hSrc == nullptr || hDst == nullptr) { HIP_RETURN(hipErrorInvalidValue); } HIP_RETURN(reinterpret_cast(hDst)->CopyAttr( @@ -2173,7 +2181,7 @@ hipError_t hipGraphDebugDotPrint(hipGraph_t graph, const char* path, unsigned in hipError_t hipGraphNodeSetEnabled(hipGraphExec_t hGraphExec, hipGraphNode_t hNode, unsigned int isEnabled) { HIP_INIT_API(hipGraphNodeSetEnabled, hGraphExec, hNode, isEnabled); - if (hGraphExec == nullptr || hNode == nullptr) { + if (hGraphExec == nullptr || hNode == nullptr) { HIP_RETURN(hipErrorInvalidValue); } hipGraphNode_t clonedNode = hGraphExec->GetClonedNode(hNode); diff --git a/projects/clr/hipamd/src/hip_graph_internal.cpp b/projects/clr/hipamd/src/hip_graph_internal.cpp index 81fb8ae337..328593630b 100644 --- a/projects/clr/hipamd/src/hip_graph_internal.cpp +++ b/projects/clr/hipamd/src/hip_graph_internal.cpp @@ -37,7 +37,7 @@ const char* GetGraphNodeTypeString(uint32_t op) { CASE_STRING(hipGraphNodeTypeWaitEvent, WaitEventNode) CASE_STRING(hipGraphNodeTypeEventRecord, EventRecordNode) CASE_STRING(hipGraphNodeTypeExtSemaphoreSignal, ExtSemaphoreSignalNode) - CASE_STRING(hipGraphNodeTypeExtSemaphoreWait, ExtSemaphoreWaitNode) + CASE_STRING(hipGraphNodeTypeExtSemaphoreWait, ExtSemaphoreWaitNode) CASE_STRING(hipGraphNodeTypeMemcpyFromSymbol, MemcpyFromSymbolNode) CASE_STRING(hipGraphNodeTypeMemcpyToSymbol, MemcpyToSymbolNode) default: @@ -73,14 +73,14 @@ hipError_t hipGraphMemcpyNode1D::ValidateParams(void* dst, const void* src, size size_t dOffset = 0; amd::Memory* dstMemory = getMemoryObject(dst, dOffset); - if ((srcMemory == nullptr) && (dstMemory != nullptr)) { //host to device + if ((srcMemory == nullptr) && (dstMemory != nullptr)) { // host to device if (origDstMemory->getContext().devices()[0] != dstMemory->getContext().devices()[0]) { return hipErrorInvalidValue; } if (kind != hipMemcpyHostToDevice) { return hipErrorInvalidValue; } - } else if ((srcMemory != nullptr) && (dstMemory == nullptr)) { //device to host + } else if ((srcMemory != nullptr) && (dstMemory == nullptr)) { // device to host if (origSrcMemory->getContext().devices()[0] != srcMemory->getContext().devices()[0]) { return hipErrorInvalidValue; } @@ -636,14 +636,12 @@ void ihipGraph::LevelOrder(std::vector& levelOrder) { } } -const ihipGraph* ihipGraph::getOriginalGraph() const { - return pOriginalGraph_; -} +const ihipGraph* ihipGraph::getOriginalGraph() const { return pOriginalGraph_; } void ihipGraph::setOriginalGraph(const ihipGraph* pOriginalGraph) { pOriginalGraph_ = pOriginalGraph; } -ihipGraph* ihipGraph::clone(std::unordered_map& clonedNodes) const{ +ihipGraph* ihipGraph::clone(std::unordered_map& clonedNodes) const { ihipGraph* newGraph = new ihipGraph(); for (auto entry : vertices_) { hipGraphNode* node = entry->clone(); @@ -674,7 +672,7 @@ ihipGraph* ihipGraph::clone(std::unordered_map& clonedNodes) const{ return newGraph; } -ihipGraph* ihipGraph::clone() const{ +ihipGraph* ihipGraph::clone() const { std::unordered_map clonedNodes; return clone(clonedNodes); } @@ -691,9 +689,9 @@ hipError_t hipGraphExec::CreateQueues(size_t numQueues) { parallelQueues_.reserve(numQueues); for (size_t i = 0; i < numQueues; i++) { amd::HostQueue* queue; - queue = new amd::HostQueue(*hip::getCurrentDevice()->asContext(), - *hip::getCurrentDevice()->devices()[0], 0, - amd::CommandQueue::RealTimeDisabled, amd::CommandQueue::Priority::Normal); + queue = new amd::HostQueue( + *hip::getCurrentDevice()->asContext(), *hip::getCurrentDevice()->devices()[0], 0, + amd::CommandQueue::RealTimeDisabled, amd::CommandQueue::Priority::Normal); bool result = (queue != nullptr) ? queue->create() : false; // Create a host queue @@ -720,8 +718,7 @@ hipError_t hipGraphExec::Init() { hipError_t FillCommands(std::vector>& parallelLists, std::unordered_map>& nodeWaitLists, - std::vector& levelOrder, - std::vector& rootCommands, + std::vector& levelOrder, std::vector& rootCommands, amd::Command*& endCommand, amd::HostQueue* queue) { hipError_t status; for (auto& node : levelOrder) { diff --git a/projects/clr/hipamd/src/hip_graph_internal.hpp b/projects/clr/hipamd/src/hip_graph_internal.hpp index 98843bfdd5..db77ed8c6d 100644 --- a/projects/clr/hipamd/src/hip_graph_internal.hpp +++ b/projects/clr/hipamd/src/hip_graph_internal.hpp @@ -36,19 +36,19 @@ typedef hipGraphNode* Node; hipError_t FillCommands(std::vector>& parallelLists, std::unordered_map>& nodeWaitLists, - std::vector& levelOrder, - std::vector& rootCommands, + std::vector& levelOrder, std::vector& rootCommands, amd::Command*& endCommand, amd::HostQueue* queue); void UpdateQueue(std::vector>& parallelLists, amd::HostQueue*& queue, hipGraphExec* ptr); struct hipUserObject : public amd::ReferenceCountedObject { - typedef void(*UserCallbackDestructor)(void* data); + typedef void (*UserCallbackDestructor)(void* data); static std::unordered_set ObjectSet_; static amd::Monitor UserObjectLock_; + public: - hipUserObject(UserCallbackDestructor callback, void* data, unsigned int flags) : ReferenceCountedObject(), - callback_(callback), data_(data), flags_(flags) { + hipUserObject(UserCallbackDestructor callback, void* data, unsigned int flags) + : ReferenceCountedObject(), callback_(callback), data_(data), flags_(flags) { amd::ScopedLock lock(UserObjectLock_); ObjectSet_.insert(this); } @@ -88,6 +88,7 @@ struct hipUserObject : public amd::ReferenceCountedObject { ObjectSet_.erase(pUsertObj); } } + private: UserCallbackDestructor callback_; void* data_; @@ -431,9 +432,7 @@ struct ihipGraph { return true; } // Delete user obj resource from graph - void RemoveUserObjGraph(hipUserObject* pUserObj) { - graphUserObj_.erase(pUserObj); - } + void RemoveUserObjGraph(hipUserObject* pUserObj) { graphUserObj_.erase(pUserObj); } // saves the original graph ptr if cloned void setOriginalGraph(const ihipGraph* pOriginalGraph); @@ -443,9 +442,8 @@ struct ihipGraph { void GetRunList(std::vector>& parallelLists, std::unordered_map>& dependencies); void LevelOrder(std::vector& levelOrder); - void GetUserObjs( std::unordered_set& graphExeUserObjs) { - for(auto userObj : graphUserObj_) - { + void GetUserObjs(std::unordered_set& graphExeUserObjs) { + for (auto userObj : graphUserObj_) { userObj->retain(); graphExeUserObjs.insert(userObj); } @@ -458,7 +456,7 @@ struct ihipGraph { for (auto node : vertices_) { node->GenerateDOTNode(GetID(), fout, flag); } - fout<<"\n"; + fout << "\n"; for (auto& node : vertices_) { node->GenerateDOTNodeEdges(GetID(), fout, flag); } @@ -485,7 +483,8 @@ struct hipGraphExec { public: hipGraphExec(std::vector& levelOrder, std::vector>& lists, std::unordered_map>& nodeWaitLists, - std::unordered_map& clonedNodes, std::unordered_set& userObjs) + std::unordered_map& clonedNodes, + std::unordered_set& userObjs) : parallelLists_(lists), levelOrder_(levelOrder), nodeWaitLists_(nodeWaitLists), @@ -713,25 +712,25 @@ class hipGraphKernelNode : public hipGraphNode { if (!func) { return hipErrorInvalidDeviceFunction; } - hip::DeviceFunc *function = hip::DeviceFunc::asFunction(func); + hip::DeviceFunc* function = hip::DeviceFunc::asFunction(func); amd::Kernel* kernel = function->kernel(); const amd::KernelSignature& signature = kernel->signature(); numParams_ = signature.numParameters(); // Allocate/assign memory if params are passed part of 'kernelParams' if (pNodeParams->kernelParams != nullptr) { - pKernelParams_->kernelParams = (void**) malloc(numParams_ * sizeof(void*)); + pKernelParams_->kernelParams = (void**)malloc(numParams_ * sizeof(void*)); if (pKernelParams_->kernelParams == nullptr) { return hipErrorOutOfMemory; } for (uint32_t i = 0; i < numParams_; ++i) { const amd::KernelParameterDescriptor& desc = signature.at(i); - pKernelParams_->kernelParams[i] = malloc(desc.size_); - if (pKernelParams_->kernelParams[i] == nullptr) { - return hipErrorOutOfMemory; - } - ::memcpy(pKernelParams_->kernelParams[i], (pNodeParams->kernelParams[i]), desc.size_); + pKernelParams_->kernelParams[i] = malloc(desc.size_); + if (pKernelParams_->kernelParams[i] == nullptr) { + return hipErrorOutOfMemory; + } + ::memcpy(pKernelParams_->kernelParams[i], (pNodeParams->kernelParams[i]), desc.size_); } } @@ -742,22 +741,22 @@ class hipGraphKernelNode : public hipGraphNode { // HIP_LAUNCH_PARAM_BUFFER_SIZE, &kernargs_size, // HIP_LAUNCH_PARAM_END } unsigned int numExtra = 5; - pKernelParams_->extra = (void**) malloc(numExtra * sizeof(void *)); + pKernelParams_->extra = (void**)malloc(numExtra * sizeof(void*)); if (pKernelParams_->extra == nullptr) { return hipErrorOutOfMemory; } pKernelParams_->extra[0] = pNodeParams->extra[0]; - size_t kernargs_size = *((size_t *)pNodeParams->extra[3]); + size_t kernargs_size = *((size_t*)pNodeParams->extra[3]); pKernelParams_->extra[1] = malloc(kernargs_size); if (pKernelParams_->extra[1] == nullptr) { return hipErrorOutOfMemory; } pKernelParams_->extra[2] = pNodeParams->extra[2]; - pKernelParams_->extra[3] = malloc(sizeof(void *)); + pKernelParams_->extra[3] = malloc(sizeof(void*)); if (pKernelParams_->extra[3] == nullptr) { return hipErrorOutOfMemory; } - *((size_t *)pKernelParams_->extra[3]) = kernargs_size; + *((size_t*)pKernelParams_->extra[3]) = kernargs_size; ::memcpy(pKernelParams_->extra[1], (pNodeParams->extra[1]), kernargs_size); pKernelParams_->extra[4] = pNodeParams->extra[4]; } @@ -773,9 +772,7 @@ class hipGraphKernelNode : public hipGraphNode { memset(&kernelAttr_, 0, sizeof(kernelAttr_)); } - ~hipGraphKernelNode() { - freeParams(); - } + ~hipGraphKernelNode() { freeParams(); } void freeParams() { // Deallocate memory allocated for kernargs passed via 'kernelParams' @@ -793,7 +790,7 @@ class hipGraphKernelNode : public hipGraphNode { else { free(pKernelParams_->extra[1]); free(pKernelParams_->extra[3]); - memset(pKernelParams_->extra, 0, 5 * sizeof(pKernelParams_->extra[0])); // 5 items + memset(pKernelParams_->extra, 0, 5 * sizeof(pKernelParams_->extra[0])); // 5 items free(pKernelParams_->extra); pKernelParams_->extra = nullptr; } @@ -805,8 +802,7 @@ class hipGraphKernelNode : public hipGraphNode { pKernelParams_ = new hipKernelNodeParams(*rhs.pKernelParams_); hipError_t status = copyParams(rhs.pKernelParams_); if (status != hipSuccess) { - ClPrint(amd::LOG_ERROR, amd::LOG_CODE, - "[hipGraph] Failed to allocate memory to copy params"); + ClPrint(amd::LOG_ERROR, amd::LOG_CODE, "[hipGraph] Failed to allocate memory to copy params"); } status = CopyAttr(&rhs); if (status != hipSuccess) { @@ -851,9 +847,9 @@ class hipGraphKernelNode : public hipGraphNode { ClPrint(amd::LOG_ERROR, amd::LOG_CODE, "[hipGraph] Failed to validateKernelParams"); return status; } - if (pKernelParams_ && ((pKernelParams_->kernelParams - && pKernelParams_->kernelParams == params->kernelParams) - || (pKernelParams_->extra && pKernelParams_->extra == params->extra))) { + if (pKernelParams_ && + ((pKernelParams_->kernelParams && pKernelParams_->kernelParams == params->kernelParams) || + (pKernelParams_->extra && pKernelParams_->extra == params->extra))) { // params is copied from pKernelParams_ and then updated, so just copy it back *pKernelParams_ = *params; return status; @@ -905,7 +901,7 @@ class hipGraphKernelNode : public hipGraphNode { return hipSuccess; } hipError_t CopyAttr(const hipGraphKernelNode* srcNode) { - if(srcNode->kernelAttrInUse_ != kernelAttrInUse_) { + if (srcNode->kernelAttrInUse_ != kernelAttrInUse_) { return hipErrorInvalidContext; } switch (kernelAttrInUse_) { @@ -914,7 +910,8 @@ class hipGraphKernelNode : public hipGraphNode { kernelAttr_.accessPolicyWindow.hitProp = srcNode->kernelAttr_.accessPolicyWindow.hitProp; kernelAttr_.accessPolicyWindow.hitRatio = srcNode->kernelAttr_.accessPolicyWindow.hitRatio; kernelAttr_.accessPolicyWindow.missProp = srcNode->kernelAttr_.accessPolicyWindow.missProp; - kernelAttr_.accessPolicyWindow.num_bytes = srcNode->kernelAttr_.accessPolicyWindow.num_bytes; + kernelAttr_.accessPolicyWindow.num_bytes = + srcNode->kernelAttr_.accessPolicyWindow.num_bytes; break; case hipKernelNodeAttributeCooperative: kernelAttr_.cooperative = srcNode->kernelAttr_.cooperative; @@ -947,7 +944,7 @@ class hipGraphKernelNode : public hipGraphNode { } static hipError_t validateKernelParams(const hipKernelNodeParams* pNodeParams, - hipFunction_t *ptrFunc = nullptr, int devId = -1) { + hipFunction_t* ptrFunc = nullptr, int devId = -1) { devId = devId == -1 ? ihipGetDevice() : devId; hipFunction_t func = getFunc(*pNodeParams, devId); if (!func) { @@ -967,10 +964,9 @@ class hipGraphKernelNode : public hipGraphNode { return status; } - if (ptrFunc) *ptrFunc = func; + if (ptrFunc) *ptrFunc = func; return hipSuccess; } - }; class hipGraphMemcpyNode : public hipGraphNode { @@ -1116,8 +1112,7 @@ class hipGraphMemcpyNode1D : public hipGraphNode { dst_(dst), src_(src), count_(count), - kind_(kind) { - } + kind_(kind) {} ~hipGraphMemcpyNode1D() {} @@ -1460,8 +1455,9 @@ class hipGraphMemsetNode : public hipGraphNode { sprintf(buffer, "{\n%s\n| {{ID | node handle | dptr | pitch | value | elementSize | width | " "height} | {%u | %p | %p | %zu | %u | %u | %zu | %zu}}}", - label_.c_str(), GetID(), this, pMemsetParams_->dst, pMemsetParams_->pitch, pMemsetParams_->value, - pMemsetParams_->elementSize, pMemsetParams_->width, pMemsetParams_->height); + label_.c_str(), GetID(), this, pMemsetParams_->dst, pMemsetParams_->pitch, + pMemsetParams_->value, pMemsetParams_->elementSize, pMemsetParams_->width, + pMemsetParams_->height); label = buffer; } else { size_t sizeBytes; @@ -1470,8 +1466,8 @@ class hipGraphMemsetNode : public hipGraphNode { } else { sizeBytes = pMemsetParams_->width * pMemsetParams_->height; } - label = std::to_string(GetID()) + "\n"+label_+"\n(" + std::to_string(pMemsetParams_->value) + - "," + std::to_string(sizeBytes) + ")"; + label = std::to_string(GetID()) + "\n" + label_ + "\n(" + + std::to_string(pMemsetParams_->value) + "," + std::to_string(sizeBytes) + ")"; } return label; } @@ -1541,7 +1537,8 @@ class hipGraphEventRecordNode : public hipGraphNode { public: hipGraphEventRecordNode(hipEvent_t event) - : hipGraphNode(hipGraphNodeTypeEventRecord, "solid", "rectangle", "EVENT_RECORD"), event_(event) {} + : hipGraphNode(hipGraphNodeTypeEventRecord, "solid", "rectangle", "EVENT_RECORD"), + event_(event) {} ~hipGraphEventRecordNode() {} hipGraphNode* clone() const { @@ -1603,7 +1600,8 @@ class hipGraphEventWaitNode : public hipGraphNode { public: hipGraphEventWaitNode(hipEvent_t event) - : hipGraphNode(hipGraphNodeTypeWaitEvent, "solid", "rectangle", "EVENT_WAIT"), event_(event) {} + : hipGraphNode(hipGraphNodeTypeWaitEvent, "solid", "rectangle", "EVENT_WAIT"), + event_(event) {} ~hipGraphEventWaitNode() {} hipGraphNode* clone() const {