diff --git a/hipamd/include/hip/amd_detail/hip_api_trace.hpp b/hipamd/include/hip/amd_detail/hip_api_trace.hpp index 850252b400..75a881e640 100644 --- a/hipamd/include/hip/amd_detail/hip_api_trace.hpp +++ b/hipamd/include/hip/amd_detail/hip_api_trace.hpp @@ -929,7 +929,7 @@ typedef hipError_t (*t_hipHccModuleLaunchKernel)(hipFunction_t f, uint32_t globa typedef int (*t_hipGetStreamDeviceId)(hipStream_t stream); typedef hipError_t (*t_hipDrvGraphAddMemsetNode)(hipGraphNode_t* phGraphNode, hipGraph_t hGraph, const hipGraphNode_t* dependencies, size_t numDependencies, - const HIP_MEMSET_NODE_PARAMS* memsetParams, hipCtx_t ctx); + const hipMemsetParams* memsetParams, hipCtx_t ctx); typedef hipError_t (*t_hipGraphAddExternalSemaphoresWaitNode)(hipGraphNode_t* pGraphNode, hipGraph_t graph, const hipGraphNode_t* pDependencies, size_t numDependencies, @@ -982,7 +982,7 @@ typedef hipError_t (*t_hipDrvGraphExecMemcpyNodeSetParams)(hipGraphExec_t hGraph hipCtx_t ctx); typedef hipError_t (*t_hipDrvGraphExecMemsetNodeSetParams)(hipGraphExec_t hGraphExec, - hipGraphNode_t hNode, const HIP_MEMSET_NODE_PARAMS* memsetParams, + hipGraphNode_t hNode, const hipMemsetParams* memsetParams, hipCtx_t ctx); typedef hipError_t (*t_hipSetValidDevices)(int* device_arr, int len); typedef hipError_t (*t_hipMemcpyAtoD)(hipDeviceptr_t dstDevice, hipArray_t srcArray, diff --git a/hipamd/include/hip/amd_detail/hip_prof_str.h b/hipamd/include/hip/amd_detail/hip_prof_str.h index 3994975ba3..9431cc4963 100644 --- a/hipamd/include/hip/amd_detail/hip_prof_str.h +++ b/hipamd/include/hip/amd_detail/hip_prof_str.h @@ -1623,8 +1623,8 @@ typedef struct hip_api_data_s { const hipGraphNode_t* dependencies; hipGraphNode_t dependencies__val; size_t numDependencies; - const HIP_MEMSET_NODE_PARAMS* memsetParams; - HIP_MEMSET_NODE_PARAMS memsetParams__val; + const hipMemsetParams* memsetParams; + hipMemsetParams memsetParams__val; hipCtx_t ctx; } hipDrvGraphAddMemsetNode; struct { @@ -1637,8 +1637,8 @@ typedef struct hip_api_data_s { struct { hipGraphExec_t hGraphExec; hipGraphNode_t hNode; - const HIP_MEMSET_NODE_PARAMS* memsetParams; - HIP_MEMSET_NODE_PARAMS memsetParams__val; + const hipMemsetParams* memsetParams; + hipMemsetParams memsetParams__val; hipCtx_t ctx; } hipDrvGraphExecMemsetNodeSetParams; struct { @@ -4082,13 +4082,13 @@ typedef struct hip_api_data_s { cb_data.args.hipDrvGraphAddMemcpyNode.copyParams = (const HIP_MEMCPY3D*)copyParams; \ cb_data.args.hipDrvGraphAddMemcpyNode.ctx = (hipCtx_t)ctx; \ }; -// hipDrvGraphAddMemsetNode[('hipGraphNode_t*', 'phGraphNode'), ('hipGraph_t', 'hGraph'), ('const hipGraphNode_t*', 'dependencies'), ('size_t', 'numDependencies'), ('const HIP_MEMSET_NODE_PARAMS*', 'memsetParams'), ('hipCtx_t', 'ctx')] +// hipDrvGraphAddMemsetNode[('hipGraphNode_t*', 'phGraphNode'), ('hipGraph_t', 'hGraph'), ('const hipGraphNode_t*', 'dependencies'), ('size_t', 'numDependencies'), ('const hipMemsetParams*', 'memsetParams'), ('hipCtx_t', 'ctx')] #define INIT_hipDrvGraphAddMemsetNode_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipDrvGraphAddMemsetNode.phGraphNode = (hipGraphNode_t*)phGraphNode; \ cb_data.args.hipDrvGraphAddMemsetNode.hGraph = (hipGraph_t)hGraph; \ cb_data.args.hipDrvGraphAddMemsetNode.dependencies = (const hipGraphNode_t*)dependencies; \ cb_data.args.hipDrvGraphAddMemsetNode.numDependencies = (size_t)numDependencies; \ - cb_data.args.hipDrvGraphAddMemsetNode.memsetParams = (const HIP_MEMSET_NODE_PARAMS*)memsetParams; \ + cb_data.args.hipDrvGraphAddMemsetNode.memsetParams = (const hipMemsetParams*)memsetParams; \ cb_data.args.hipDrvGraphAddMemsetNode.ctx = (hipCtx_t)ctx; \ }; // hipDrvGraphExecMemcpyNodeSetParams[('hipGraphExec_t', 'hGraphExec'), ('hipGraphNode_t', 'hNode'), ('const HIP_MEMCPY3D*', 'copyParams'), ('hipCtx_t', 'ctx')] @@ -4098,11 +4098,11 @@ typedef struct hip_api_data_s { cb_data.args.hipDrvGraphExecMemcpyNodeSetParams.copyParams = (const HIP_MEMCPY3D*)copyParams; \ cb_data.args.hipDrvGraphExecMemcpyNodeSetParams.ctx = (hipCtx_t)ctx; \ }; -// hipDrvGraphExecMemsetNodeSetParams[('hipGraphExec_t', 'hGraphExec'), ('hipGraphNode_t', 'hNode'), ('const HIP_MEMSET_NODE_PARAMS*', 'memsetParams'), ('hipCtx_t', 'ctx')] +// hipDrvGraphExecMemsetNodeSetParams[('hipGraphExec_t', 'hGraphExec'), ('hipGraphNode_t', 'hNode'), ('const hipMemsetParams*', 'memsetParams'), ('hipCtx_t', 'ctx')] #define INIT_hipDrvGraphExecMemsetNodeSetParams_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipDrvGraphExecMemsetNodeSetParams.hGraphExec = (hipGraphExec_t)hGraphExec; \ cb_data.args.hipDrvGraphExecMemsetNodeSetParams.hNode = (hipGraphNode_t)hNode; \ - cb_data.args.hipDrvGraphExecMemsetNodeSetParams.memsetParams = (const HIP_MEMSET_NODE_PARAMS*)memsetParams; \ + cb_data.args.hipDrvGraphExecMemsetNodeSetParams.memsetParams = (const hipMemsetParams*)memsetParams; \ cb_data.args.hipDrvGraphExecMemsetNodeSetParams.ctx = (hipCtx_t)ctx; \ }; // hipDrvGraphMemcpyNodeGetParams[('hipGraphNode_t', 'hNode'), ('HIP_MEMCPY3D*', 'nodeParams')] @@ -6551,7 +6551,7 @@ static inline void hipApiArgsInit(hip_api_id_t id, hip_api_data_t* data) { if (data->args.hipDrvGraphAddMemcpyNode.dependencies) data->args.hipDrvGraphAddMemcpyNode.dependencies__val = *(data->args.hipDrvGraphAddMemcpyNode.dependencies); if (data->args.hipDrvGraphAddMemcpyNode.copyParams) data->args.hipDrvGraphAddMemcpyNode.copyParams__val = *(data->args.hipDrvGraphAddMemcpyNode.copyParams); break; -// hipDrvGraphAddMemsetNode[('hipGraphNode_t*', 'phGraphNode'), ('hipGraph_t', 'hGraph'), ('const hipGraphNode_t*', 'dependencies'), ('size_t', 'numDependencies'), ('const HIP_MEMSET_NODE_PARAMS*', 'memsetParams'), ('hipCtx_t', 'ctx')] +// hipDrvGraphAddMemsetNode[('hipGraphNode_t*', 'phGraphNode'), ('hipGraph_t', 'hGraph'), ('const hipGraphNode_t*', 'dependencies'), ('size_t', 'numDependencies'), ('const hipMemsetParams*', 'memsetParams'), ('hipCtx_t', 'ctx')] case HIP_API_ID_hipDrvGraphAddMemsetNode: if (data->args.hipDrvGraphAddMemsetNode.phGraphNode) data->args.hipDrvGraphAddMemsetNode.phGraphNode__val = *(data->args.hipDrvGraphAddMemsetNode.phGraphNode); if (data->args.hipDrvGraphAddMemsetNode.dependencies) data->args.hipDrvGraphAddMemsetNode.dependencies__val = *(data->args.hipDrvGraphAddMemsetNode.dependencies); @@ -6561,7 +6561,7 @@ static inline void hipApiArgsInit(hip_api_id_t id, hip_api_data_t* data) { case HIP_API_ID_hipDrvGraphExecMemcpyNodeSetParams: if (data->args.hipDrvGraphExecMemcpyNodeSetParams.copyParams) data->args.hipDrvGraphExecMemcpyNodeSetParams.copyParams__val = *(data->args.hipDrvGraphExecMemcpyNodeSetParams.copyParams); break; -// hipDrvGraphExecMemsetNodeSetParams[('hipGraphExec_t', 'hGraphExec'), ('hipGraphNode_t', 'hNode'), ('const HIP_MEMSET_NODE_PARAMS*', 'memsetParams'), ('hipCtx_t', 'ctx')] +// hipDrvGraphExecMemsetNodeSetParams[('hipGraphExec_t', 'hGraphExec'), ('hipGraphNode_t', 'hNode'), ('const hipMemsetParams*', 'memsetParams'), ('hipCtx_t', 'ctx')] case HIP_API_ID_hipDrvGraphExecMemsetNodeSetParams: if (data->args.hipDrvGraphExecMemsetNodeSetParams.memsetParams) data->args.hipDrvGraphExecMemsetNodeSetParams.memsetParams__val = *(data->args.hipDrvGraphExecMemsetNodeSetParams.memsetParams); break; diff --git a/hipamd/src/hip_api_trace.cpp b/hipamd/src/hip_api_trace.cpp index 59992f5241..23c31799d6 100644 --- a/hipamd/src/hip_api_trace.cpp +++ b/hipamd/src/hip_api_trace.cpp @@ -359,7 +359,7 @@ hipError_t hipImportExternalSemaphore(hipExternalSemaphore_t* extSem_out, const hipExternalSemaphoreHandleDesc* semHandleDesc); hipError_t hipDrvGraphAddMemsetNode(hipGraphNode_t* phGraphNode, hipGraph_t hGraph, const hipGraphNode_t* dependencies, size_t numDependencies, - const HIP_MEMSET_NODE_PARAMS* memsetParams, hipCtx_t ctx); + const hipMemsetParams* memsetParams, hipCtx_t ctx); hipError_t hipInit(unsigned int flags); hipError_t hipIpcCloseMemHandle(void* devPtr); hipError_t hipIpcGetEventHandle(hipIpcEventHandle_t* handle, hipEvent_t event); @@ -789,7 +789,7 @@ hipError_t hipDrvGraphAddMemFreeNode(hipGraphNode_t* phGraphNode, hipGraph_t hGr hipError_t hipDrvGraphExecMemcpyNodeSetParams(hipGraphExec_t hGraphExec, hipGraphNode_t hNode, const HIP_MEMCPY3D* copyParams, hipCtx_t ctx); hipError_t hipDrvGraphExecMemsetNodeSetParams(hipGraphExec_t hGraphExec, hipGraphNode_t hNode, - const HIP_MEMSET_NODE_PARAMS* memsetParams, hipCtx_t ctx); + const hipMemsetParams* memsetParams, hipCtx_t ctx); hipError_t hipSetValidDevices(int* device_arr, int len); hipError_t hipMemcpyAtoD(hipDeviceptr_t dstDevice, hipArray_t srcArray, size_t srcOffset, size_t ByteCount); diff --git a/hipamd/src/hip_graph.cpp b/hipamd/src/hip_graph.cpp index 6835f62ec2..1f45881df2 100644 --- a/hipamd/src/hip_graph.cpp +++ b/hipamd/src/hip_graph.cpp @@ -1411,7 +1411,7 @@ hipError_t hipGraphAddMemsetNode(hipGraphNode_t* pGraphNode, hipGraph_t graph, hipError_t hipDrvGraphAddMemsetNode(hipGraphNode_t* phGraphNode, hipGraph_t hGraph, const hipGraphNode_t* dependencies, size_t numDependencies, - const HIP_MEMSET_NODE_PARAMS* memsetParams, hipCtx_t ctx) { + const hipMemsetParams* memsetParams, hipCtx_t ctx) { HIP_INIT_API(hipDrvGraphAddMemsetNode, phGraphNode, hGraph, dependencies, numDependencies, memsetParams, ctx); if (phGraphNode == nullptr || hGraph == nullptr || @@ -3393,7 +3393,7 @@ hipError_t hipDrvGraphExecMemcpyNodeSetParams(hipGraphExec_t hGraphExec, hipGrap } hipError_t hipDrvGraphExecMemsetNodeSetParams(hipGraphExec_t hGraphExec, hipGraphNode_t hNode, - const HIP_MEMSET_NODE_PARAMS* memsetParams, hipCtx_t ctx) { + const hipMemsetParams* memsetParams, hipCtx_t ctx) { HIP_INIT_API(hipDrvGraphExecMemsetNodeSetParams, hGraphExec, hNode, memsetParams); hip::GraphNode* n = reinterpret_cast(hNode); diff --git a/hipamd/src/hip_graph_internal.hpp b/hipamd/src/hip_graph_internal.hpp index 56084cc17f..51b75f59f4 100644 --- a/hipamd/src/hip_graph_internal.hpp +++ b/hipamd/src/hip_graph_internal.hpp @@ -2005,15 +2005,6 @@ class GraphMemsetNode : public GraphNode { std::memcpy(params, &memsetParams_, sizeof(hipMemsetParams)); } - void GetParams(HIP_MEMSET_NODE_PARAMS* params) { - params->dst = memsetParams_.dst; - params->elementSize = memsetParams_.elementSize; - params->height = memsetParams_.height; - params->pitch = memsetParams_.pitch; - params->value = memsetParams_.value; - params->width = memsetParams_.width; - } - hipError_t SetParamsInternal(const hipMemsetParams* params, bool isExec, size_t depth = 1) { hipError_t hip_error = hipSuccess; hip_error = ihipGraphMemsetParams_validate(params); @@ -2086,16 +2077,6 @@ class GraphMemsetNode : public GraphNode { return SetParamsInternal(params, isExec, depth); } - hipError_t SetParams(const HIP_MEMSET_NODE_PARAMS* params, bool isExec = false, size_t depth = 1) { - hipMemsetParams pmemsetParams; - pmemsetParams.dst = params->dst; - pmemsetParams.elementSize = params->elementSize; - pmemsetParams.height = params->height; - pmemsetParams.pitch = params->pitch; - pmemsetParams.value = params->value; - pmemsetParams.width = params->width; - return SetParamsInternal(&pmemsetParams, isExec, depth); - } hipError_t SetParams(GraphNode* node) override { const GraphMemsetNode* memsetNode = static_cast(node); return SetParams(&memsetNode->memsetParams_, false, memsetNode->depth_); diff --git a/hipamd/src/hip_table_interface.cpp b/hipamd/src/hip_table_interface.cpp index 7749c1f233..830e17f741 100644 --- a/hipamd/src/hip_table_interface.cpp +++ b/hipamd/src/hip_table_interface.cpp @@ -831,7 +831,7 @@ hipError_t hipImportExternalSemaphore(hipExternalSemaphore_t* extSem_out, } hipError_t hipDrvGraphAddMemsetNode(hipGraphNode_t* phGraphNode, hipGraph_t hGraph, const hipGraphNode_t* dependencies, size_t numDependencies, - const HIP_MEMSET_NODE_PARAMS* memsetParams, hipCtx_t ctx) { + const hipMemsetParams* memsetParams, hipCtx_t ctx) { return hip::GetHipDispatchTable()->hipDrvGraphAddMemsetNode_fn(phGraphNode, hGraph, dependencies, numDependencies, memsetParams, ctx); } @@ -1785,7 +1785,7 @@ hipError_t hipGetFuncBySymbol(hipFunction_t* functionPtr, const void* symbolPtr) return hip::GetHipDispatchTable()->hipGetFuncBySymbol_fn(functionPtr, symbolPtr); } hipError_t hipDrvGraphExecMemsetNodeSetParams(hipGraphExec_t hGraphExec, hipGraphNode_t hNode, - const HIP_MEMSET_NODE_PARAMS* memsetParams, hipCtx_t ctx) { + const hipMemsetParams* memsetParams, hipCtx_t ctx) { return hip::GetHipDispatchTable()->hipDrvGraphExecMemsetNodeSetParams_fn(hGraphExec, hNode, memsetParams, ctx); }