diff --git a/hipamd/include/hip/amd_detail/hip_prof_str.h b/hipamd/include/hip/amd_detail/hip_prof_str.h index d3414ec0c4..c502cf8be5 100644 --- a/hipamd/include/hip/amd_detail/hip_prof_str.h +++ b/hipamd/include/hip/amd_detail/hip_prof_str.h @@ -306,9 +306,14 @@ enum hip_api_id_t { HIP_API_ID_hipGraphicsGLRegisterImage = 293, HIP_API_ID_hipGraphicsSubResourceGetMappedArray = 294, HIP_API_ID_hipPointerGetAttribute = 295, - HIP_API_ID_hipTexRefSetArray = 296, + HIP_API_ID_RESERVED_296 = 296, HIP_API_ID_hipThreadExchangeStreamCaptureMode = 297, - HIP_API_ID_LAST = 297, + HIP_API_ID_hipDeviceGetUuid = 298, + HIP_API_ID_hipGetChannelDesc = 299, + HIP_API_ID_hipGraphKernelNodeGetAttribute = 300, + HIP_API_ID_hipGraphKernelNodeSetAttribute = 301, + HIP_API_ID_hipLaunchHostFunc = 302, + HIP_API_ID_LAST = 302, HIP_API_ID_hipArray3DGetDescriptor = HIP_API_ID_NONE, HIP_API_ID_hipArrayGetDescriptor = HIP_API_ID_NONE, @@ -319,7 +324,6 @@ enum hip_api_id_t { HIP_API_ID_hipCreateTextureObject = HIP_API_ID_NONE, HIP_API_ID_hipDestroyTextureObject = HIP_API_ID_NONE, HIP_API_ID_hipDeviceGetCount = HIP_API_ID_NONE, - HIP_API_ID_hipGetChannelDesc = HIP_API_ID_NONE, HIP_API_ID_hipGetTextureAlignmentOffset = HIP_API_ID_NONE, HIP_API_ID_hipGetTextureObjectResourceDesc = HIP_API_ID_NONE, HIP_API_ID_hipGetTextureObjectResourceViewDesc = HIP_API_ID_NONE, @@ -348,6 +352,7 @@ enum hip_api_id_t { HIP_API_ID_hipTexRefGetMipmapFilterMode = HIP_API_ID_NONE, HIP_API_ID_hipTexRefGetMipmappedArray = HIP_API_ID_NONE, HIP_API_ID_hipTexRefSetAddressMode = HIP_API_ID_NONE, + HIP_API_ID_hipTexRefSetArray = HIP_API_ID_NONE, HIP_API_ID_hipTexRefSetFilterMode = HIP_API_ID_NONE, HIP_API_ID_hipTexRefSetFlags = HIP_API_ID_NONE, HIP_API_ID_hipTexRefSetMipmapFilterMode = HIP_API_ID_NONE, @@ -400,6 +405,7 @@ static inline const char* hip_api_name(const uint32_t id) { case HIP_API_ID_hipDeviceGetPCIBusId: return "hipDeviceGetPCIBusId"; case HIP_API_ID_hipDeviceGetSharedMemConfig: return "hipDeviceGetSharedMemConfig"; case HIP_API_ID_hipDeviceGetStreamPriorityRange: return "hipDeviceGetStreamPriorityRange"; + case HIP_API_ID_hipDeviceGetUuid: return "hipDeviceGetUuid"; case HIP_API_ID_hipDevicePrimaryCtxGetState: return "hipDevicePrimaryCtxGetState"; case HIP_API_ID_hipDevicePrimaryCtxRelease: return "hipDevicePrimaryCtxRelease"; case HIP_API_ID_hipDevicePrimaryCtxReset: return "hipDevicePrimaryCtxReset"; @@ -439,6 +445,7 @@ static inline const char* hip_api_name(const uint32_t id) { case HIP_API_ID_hipFuncSetCacheConfig: return "hipFuncSetCacheConfig"; case HIP_API_ID_hipFuncSetSharedMemConfig: return "hipFuncSetSharedMemConfig"; case HIP_API_ID_hipGLGetDevices: return "hipGLGetDevices"; + case HIP_API_ID_hipGetChannelDesc: return "hipGetChannelDesc"; case HIP_API_ID_hipGetDevice: return "hipGetDevice"; case HIP_API_ID_hipGetDeviceCount: return "hipGetDeviceCount"; case HIP_API_ID_hipGetDeviceFlags: return "hipGetDeviceFlags"; @@ -488,7 +495,9 @@ static inline const char* hip_api_name(const uint32_t id) { case HIP_API_ID_hipGraphHostNodeSetParams: return "hipGraphHostNodeSetParams"; case HIP_API_ID_hipGraphInstantiate: return "hipGraphInstantiate"; case HIP_API_ID_hipGraphInstantiateWithFlags: return "hipGraphInstantiateWithFlags"; + case HIP_API_ID_hipGraphKernelNodeGetAttribute: return "hipGraphKernelNodeGetAttribute"; case HIP_API_ID_hipGraphKernelNodeGetParams: return "hipGraphKernelNodeGetParams"; + case HIP_API_ID_hipGraphKernelNodeSetAttribute: return "hipGraphKernelNodeSetAttribute"; case HIP_API_ID_hipGraphKernelNodeSetParams: return "hipGraphKernelNodeSetParams"; case HIP_API_ID_hipGraphLaunch: return "hipGraphLaunch"; case HIP_API_ID_hipGraphMemcpyNodeGetParams: return "hipGraphMemcpyNodeGetParams"; @@ -529,6 +538,7 @@ static inline const char* hip_api_name(const uint32_t id) { case HIP_API_ID_hipLaunchByPtr: return "hipLaunchByPtr"; case HIP_API_ID_hipLaunchCooperativeKernel: return "hipLaunchCooperativeKernel"; case HIP_API_ID_hipLaunchCooperativeKernelMultiDevice: return "hipLaunchCooperativeKernelMultiDevice"; + case HIP_API_ID_hipLaunchHostFunc: return "hipLaunchHostFunc"; case HIP_API_ID_hipLaunchKernel: return "hipLaunchKernel"; case HIP_API_ID_hipMalloc: return "hipMalloc"; case HIP_API_ID_hipMalloc3D: return "hipMalloc3D"; @@ -645,7 +655,6 @@ static inline const char* hip_api_name(const uint32_t id) { case HIP_API_ID_hipTexRefGetMipmapLevelClamp: return "hipTexRefGetMipmapLevelClamp"; case HIP_API_ID_hipTexRefSetAddress: return "hipTexRefSetAddress"; case HIP_API_ID_hipTexRefSetAddress2D: return "hipTexRefSetAddress2D"; - case HIP_API_ID_hipTexRefSetArray: return "hipTexRefSetArray"; case HIP_API_ID_hipTexRefSetBorderColor: return "hipTexRefSetBorderColor"; case HIP_API_ID_hipTexRefSetFormat: return "hipTexRefSetFormat"; case HIP_API_ID_hipTexRefSetMaxAnisotropy: return "hipTexRefSetMaxAnisotropy"; @@ -701,6 +710,7 @@ static inline uint32_t hipApiIdByName(const char* name) { if (strcmp("hipDeviceGetPCIBusId", name) == 0) return HIP_API_ID_hipDeviceGetPCIBusId; if (strcmp("hipDeviceGetSharedMemConfig", name) == 0) return HIP_API_ID_hipDeviceGetSharedMemConfig; if (strcmp("hipDeviceGetStreamPriorityRange", name) == 0) return HIP_API_ID_hipDeviceGetStreamPriorityRange; + if (strcmp("hipDeviceGetUuid", name) == 0) return HIP_API_ID_hipDeviceGetUuid; if (strcmp("hipDevicePrimaryCtxGetState", name) == 0) return HIP_API_ID_hipDevicePrimaryCtxGetState; if (strcmp("hipDevicePrimaryCtxRelease", name) == 0) return HIP_API_ID_hipDevicePrimaryCtxRelease; if (strcmp("hipDevicePrimaryCtxReset", name) == 0) return HIP_API_ID_hipDevicePrimaryCtxReset; @@ -740,6 +750,7 @@ static inline uint32_t hipApiIdByName(const char* name) { if (strcmp("hipFuncSetCacheConfig", name) == 0) return HIP_API_ID_hipFuncSetCacheConfig; if (strcmp("hipFuncSetSharedMemConfig", name) == 0) return HIP_API_ID_hipFuncSetSharedMemConfig; if (strcmp("hipGLGetDevices", name) == 0) return HIP_API_ID_hipGLGetDevices; + if (strcmp("hipGetChannelDesc", name) == 0) return HIP_API_ID_hipGetChannelDesc; if (strcmp("hipGetDevice", name) == 0) return HIP_API_ID_hipGetDevice; if (strcmp("hipGetDeviceCount", name) == 0) return HIP_API_ID_hipGetDeviceCount; if (strcmp("hipGetDeviceFlags", name) == 0) return HIP_API_ID_hipGetDeviceFlags; @@ -789,7 +800,9 @@ static inline uint32_t hipApiIdByName(const char* name) { if (strcmp("hipGraphHostNodeSetParams", name) == 0) return HIP_API_ID_hipGraphHostNodeSetParams; if (strcmp("hipGraphInstantiate", name) == 0) return HIP_API_ID_hipGraphInstantiate; if (strcmp("hipGraphInstantiateWithFlags", name) == 0) return HIP_API_ID_hipGraphInstantiateWithFlags; + if (strcmp("hipGraphKernelNodeGetAttribute", name) == 0) return HIP_API_ID_hipGraphKernelNodeGetAttribute; if (strcmp("hipGraphKernelNodeGetParams", name) == 0) return HIP_API_ID_hipGraphKernelNodeGetParams; + if (strcmp("hipGraphKernelNodeSetAttribute", name) == 0) return HIP_API_ID_hipGraphKernelNodeSetAttribute; if (strcmp("hipGraphKernelNodeSetParams", name) == 0) return HIP_API_ID_hipGraphKernelNodeSetParams; if (strcmp("hipGraphLaunch", name) == 0) return HIP_API_ID_hipGraphLaunch; if (strcmp("hipGraphMemcpyNodeGetParams", name) == 0) return HIP_API_ID_hipGraphMemcpyNodeGetParams; @@ -830,6 +843,7 @@ static inline uint32_t hipApiIdByName(const char* name) { if (strcmp("hipLaunchByPtr", name) == 0) return HIP_API_ID_hipLaunchByPtr; if (strcmp("hipLaunchCooperativeKernel", name) == 0) return HIP_API_ID_hipLaunchCooperativeKernel; if (strcmp("hipLaunchCooperativeKernelMultiDevice", name) == 0) return HIP_API_ID_hipLaunchCooperativeKernelMultiDevice; + if (strcmp("hipLaunchHostFunc", name) == 0) return HIP_API_ID_hipLaunchHostFunc; if (strcmp("hipLaunchKernel", name) == 0) return HIP_API_ID_hipLaunchKernel; if (strcmp("hipMalloc", name) == 0) return HIP_API_ID_hipMalloc; if (strcmp("hipMalloc3D", name) == 0) return HIP_API_ID_hipMalloc3D; @@ -946,7 +960,6 @@ static inline uint32_t hipApiIdByName(const char* name) { if (strcmp("hipTexRefGetMipmapLevelClamp", name) == 0) return HIP_API_ID_hipTexRefGetMipmapLevelClamp; if (strcmp("hipTexRefSetAddress", name) == 0) return HIP_API_ID_hipTexRefSetAddress; if (strcmp("hipTexRefSetAddress2D", name) == 0) return HIP_API_ID_hipTexRefSetAddress2D; - if (strcmp("hipTexRefSetArray", name) == 0) return HIP_API_ID_hipTexRefSetArray; if (strcmp("hipTexRefSetBorderColor", name) == 0) return HIP_API_ID_hipTexRefSetBorderColor; if (strcmp("hipTexRefSetFormat", name) == 0) return HIP_API_ID_hipTexRefSetFormat; if (strcmp("hipTexRefSetMaxAnisotropy", name) == 0) return HIP_API_ID_hipTexRefSetMaxAnisotropy; @@ -1153,6 +1166,11 @@ typedef struct hip_api_data_s { int* greatestPriority; int greatestPriority__val; } hipDeviceGetStreamPriorityRange; + struct { + hipUUID* uuid; + hipUUID uuid__val; + hipDevice_t device; + } hipDeviceGetUuid; struct { hipDevice_t dev; unsigned int* flags; @@ -1350,6 +1368,11 @@ typedef struct hip_api_data_s { unsigned int hipDeviceCount; hipGLDeviceList deviceList; } hipGLGetDevices; + struct { + hipChannelFormatDesc* desc; + hipChannelFormatDesc desc__val; + hipArray_const_t array; + } hipGetChannelDesc; struct { int* deviceId; int deviceId__val; @@ -1668,11 +1691,23 @@ typedef struct hip_api_data_s { hipGraph_t graph; unsigned long long flags; } hipGraphInstantiateWithFlags; + struct { + hipGraphNode_t hNode; + hipKernelNodeAttrID attr; + hipKernelNodeAttrValue* value; + hipKernelNodeAttrValue value__val; + } hipGraphKernelNodeGetAttribute; struct { hipGraphNode_t node; hipKernelNodeParams* pNodeParams; hipKernelNodeParams pNodeParams__val; } hipGraphKernelNodeGetParams; + struct { + hipGraphNode_t hNode; + hipKernelNodeAttrID attr; + const hipKernelNodeAttrValue* value; + hipKernelNodeAttrValue value__val; + } hipGraphKernelNodeSetAttribute; struct { hipGraphNode_t node; const hipKernelNodeParams* pNodeParams; @@ -1908,6 +1943,11 @@ typedef struct hip_api_data_s { int numDevices; unsigned int flags; } hipLaunchCooperativeKernelMultiDevice; + struct { + hipStream_t stream; + hipHostFn_t fn; + void* userData; + } hipLaunchHostFunc; struct { const void* function_address; dim3 numBlocks; @@ -2659,12 +2699,6 @@ typedef struct hip_api_data_s { hipDeviceptr_t dptr; size_t Pitch; } hipTexRefSetAddress2D; - struct { - textureReference* tex; - textureReference tex__val; - hipArray_const_t array; - unsigned int flags; - } hipTexRefSetArray; struct { textureReference* texRef; textureReference texRef__val; @@ -2909,6 +2943,11 @@ typedef struct hip_api_data_s { cb_data.args.hipDeviceGetStreamPriorityRange.leastPriority = (int*)leastPriority; \ cb_data.args.hipDeviceGetStreamPriorityRange.greatestPriority = (int*)greatestPriority; \ }; +// hipDeviceGetUuid[('hipUUID*', 'uuid'), ('hipDevice_t', 'device')] +#define INIT_hipDeviceGetUuid_CB_ARGS_DATA(cb_data) { \ + cb_data.args.hipDeviceGetUuid.uuid = (hipUUID*)uuid; \ + cb_data.args.hipDeviceGetUuid.device = (hipDevice_t)device; \ +}; // hipDevicePrimaryCtxGetState[('hipDevice_t', 'dev'), ('unsigned int*', 'flags'), ('int*', 'active')] #define INIT_hipDevicePrimaryCtxGetState_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipDevicePrimaryCtxGetState.dev = (hipDevice_t)dev; \ @@ -3120,6 +3159,11 @@ typedef struct hip_api_data_s { cb_data.args.hipGLGetDevices.hipDeviceCount = (unsigned int)hipDeviceCount; \ cb_data.args.hipGLGetDevices.deviceList = (hipGLDeviceList)deviceList; \ }; +// hipGetChannelDesc[('hipChannelFormatDesc*', 'desc'), ('hipArray_const_t', 'array')] +#define INIT_hipGetChannelDesc_CB_ARGS_DATA(cb_data) { \ + cb_data.args.hipGetChannelDesc.desc = (hipChannelFormatDesc*)desc; \ + cb_data.args.hipGetChannelDesc.array = (hipArray_const_t)array; \ +}; // hipGetDevice[('int*', 'deviceId')] #define INIT_hipGetDevice_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipGetDevice.deviceId = (int*)deviceId; \ @@ -3432,11 +3476,17 @@ typedef struct hip_api_data_s { cb_data.args.hipGraphInstantiateWithFlags.graph = (hipGraph_t)graph; \ cb_data.args.hipGraphInstantiateWithFlags.flags = (unsigned long long)flags; \ }; +// hipGraphKernelNodeGetAttribute[('hipGraphNode_t', 'hNode'), ('hipKernelNodeAttrID', 'attr'), ('hipKernelNodeAttrValue*', 'value')] +#define INIT_hipGraphKernelNodeGetAttribute_CB_ARGS_DATA(cb_data) { \ +}; // hipGraphKernelNodeGetParams[('hipGraphNode_t', 'node'), ('hipKernelNodeParams*', 'pNodeParams')] #define INIT_hipGraphKernelNodeGetParams_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipGraphKernelNodeGetParams.node = (hipGraphNode_t)node; \ cb_data.args.hipGraphKernelNodeGetParams.pNodeParams = (hipKernelNodeParams*)pNodeParams; \ }; +// hipGraphKernelNodeSetAttribute[('hipGraphNode_t', 'hNode'), ('hipKernelNodeAttrID', 'attr'), ('const hipKernelNodeAttrValue*', 'value')] +#define INIT_hipGraphKernelNodeSetAttribute_CB_ARGS_DATA(cb_data) { \ +}; // hipGraphKernelNodeSetParams[('hipGraphNode_t', 'node'), ('const hipKernelNodeParams*', 'pNodeParams')] #define INIT_hipGraphKernelNodeSetParams_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipGraphKernelNodeSetParams.node = (hipGraphNode_t)node; \ @@ -3676,6 +3726,9 @@ typedef struct hip_api_data_s { cb_data.args.hipLaunchCooperativeKernelMultiDevice.numDevices = (int)numDevices; \ cb_data.args.hipLaunchCooperativeKernelMultiDevice.flags = (unsigned int)flags; \ }; +// hipLaunchHostFunc[('hipStream_t', 'stream'), ('hipHostFn_t', 'fn'), ('void*', 'userData')] +#define INIT_hipLaunchHostFunc_CB_ARGS_DATA(cb_data) { \ +}; // hipLaunchKernel[('const void*', 'function_address'), ('dim3', 'numBlocks'), ('dim3', 'dimBlocks'), ('void**', 'args'), ('size_t', 'sharedMemBytes'), ('hipStream_t', 'stream')] #define INIT_hipLaunchKernel_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipLaunchKernel.function_address = (const void*)hostFunction; \ @@ -4448,12 +4501,6 @@ typedef struct hip_api_data_s { cb_data.args.hipTexRefSetAddress2D.dptr = (hipDeviceptr_t)dptr; \ cb_data.args.hipTexRefSetAddress2D.Pitch = (size_t)Pitch; \ }; -// hipTexRefSetArray[('textureReference*', 'tex'), ('hipArray_const_t', 'array'), ('unsigned int', 'flags')] -#define INIT_hipTexRefSetArray_CB_ARGS_DATA(cb_data) { \ - cb_data.args.hipTexRefSetArray.tex = (textureReference*)texRef; \ - cb_data.args.hipTexRefSetArray.array = (hipArray_const_t)array; \ - cb_data.args.hipTexRefSetArray.flags = (unsigned int)flags; \ -}; // hipTexRefSetBorderColor[('textureReference*', 'texRef'), ('float*', 'pBorderColor')] #define INIT_hipTexRefSetBorderColor_CB_ARGS_DATA(cb_data) { \ cb_data.args.hipTexRefSetBorderColor.texRef = (textureReference*)texRef; \ @@ -4514,8 +4561,6 @@ typedef struct hip_api_data_s { #define INIT_hipDestroyTextureObject_CB_ARGS_DATA(cb_data) {}; // hipDeviceGetCount() #define INIT_hipDeviceGetCount_CB_ARGS_DATA(cb_data) {}; -// hipGetChannelDesc() -#define INIT_hipGetChannelDesc_CB_ARGS_DATA(cb_data) {}; // hipGetTextureAlignmentOffset() #define INIT_hipGetTextureAlignmentOffset_CB_ARGS_DATA(cb_data) {}; // hipGetTextureObjectResourceDesc() @@ -4572,6 +4617,8 @@ typedef struct hip_api_data_s { #define INIT_hipTexRefGetMipmappedArray_CB_ARGS_DATA(cb_data) {}; // hipTexRefSetAddressMode() #define INIT_hipTexRefSetAddressMode_CB_ARGS_DATA(cb_data) {}; +// hipTexRefSetArray() +#define INIT_hipTexRefSetArray_CB_ARGS_DATA(cb_data) {}; // hipTexRefSetFilterMode() #define INIT_hipTexRefSetFilterMode_CB_ARGS_DATA(cb_data) {}; // hipTexRefSetFlags() @@ -4750,6 +4797,10 @@ static inline void hipApiArgsInit(hip_api_id_t id, hip_api_data_t* data) { if (data->args.hipDeviceGetStreamPriorityRange.leastPriority) data->args.hipDeviceGetStreamPriorityRange.leastPriority__val = *(data->args.hipDeviceGetStreamPriorityRange.leastPriority); if (data->args.hipDeviceGetStreamPriorityRange.greatestPriority) data->args.hipDeviceGetStreamPriorityRange.greatestPriority__val = *(data->args.hipDeviceGetStreamPriorityRange.greatestPriority); break; +// hipDeviceGetUuid[('hipUUID*', 'uuid'), ('hipDevice_t', 'device')] + case HIP_API_ID_hipDeviceGetUuid: + if (data->args.hipDeviceGetUuid.uuid) data->args.hipDeviceGetUuid.uuid__val = *(data->args.hipDeviceGetUuid.uuid); + break; // hipDevicePrimaryCtxGetState[('hipDevice_t', 'dev'), ('unsigned int*', 'flags'), ('int*', 'active')] case HIP_API_ID_hipDevicePrimaryCtxGetState: if (data->args.hipDevicePrimaryCtxGetState.flags) data->args.hipDevicePrimaryCtxGetState.flags__val = *(data->args.hipDevicePrimaryCtxGetState.flags); @@ -4896,6 +4947,10 @@ static inline void hipApiArgsInit(hip_api_id_t id, hip_api_data_t* data) { if (data->args.hipGLGetDevices.pHipDeviceCount) data->args.hipGLGetDevices.pHipDeviceCount__val = *(data->args.hipGLGetDevices.pHipDeviceCount); if (data->args.hipGLGetDevices.pHipDevices) data->args.hipGLGetDevices.pHipDevices__val = *(data->args.hipGLGetDevices.pHipDevices); break; +// hipGetChannelDesc[('hipChannelFormatDesc*', 'desc'), ('hipArray_const_t', 'array')] + case HIP_API_ID_hipGetChannelDesc: + if (data->args.hipGetChannelDesc.desc) data->args.hipGetChannelDesc.desc__val = *(data->args.hipGetChannelDesc.desc); + break; // hipGetDevice[('int*', 'deviceId')] case HIP_API_ID_hipGetDevice: if (data->args.hipGetDevice.deviceId) data->args.hipGetDevice.deviceId__val = *(data->args.hipGetDevice.deviceId); @@ -5102,10 +5157,18 @@ static inline void hipApiArgsInit(hip_api_id_t id, hip_api_data_t* data) { case HIP_API_ID_hipGraphInstantiateWithFlags: if (data->args.hipGraphInstantiateWithFlags.pGraphExec) data->args.hipGraphInstantiateWithFlags.pGraphExec__val = *(data->args.hipGraphInstantiateWithFlags.pGraphExec); break; +// hipGraphKernelNodeGetAttribute[('hipGraphNode_t', 'hNode'), ('hipKernelNodeAttrID', 'attr'), ('hipKernelNodeAttrValue*', 'value')] + case HIP_API_ID_hipGraphKernelNodeGetAttribute: + if (data->args.hipGraphKernelNodeGetAttribute.value) data->args.hipGraphKernelNodeGetAttribute.value__val = *(data->args.hipGraphKernelNodeGetAttribute.value); + break; // hipGraphKernelNodeGetParams[('hipGraphNode_t', 'node'), ('hipKernelNodeParams*', 'pNodeParams')] case HIP_API_ID_hipGraphKernelNodeGetParams: if (data->args.hipGraphKernelNodeGetParams.pNodeParams) data->args.hipGraphKernelNodeGetParams.pNodeParams__val = *(data->args.hipGraphKernelNodeGetParams.pNodeParams); break; +// hipGraphKernelNodeSetAttribute[('hipGraphNode_t', 'hNode'), ('hipKernelNodeAttrID', 'attr'), ('const hipKernelNodeAttrValue*', 'value')] + case HIP_API_ID_hipGraphKernelNodeSetAttribute: + if (data->args.hipGraphKernelNodeSetAttribute.value) data->args.hipGraphKernelNodeSetAttribute.value__val = *(data->args.hipGraphKernelNodeSetAttribute.value); + break; // hipGraphKernelNodeSetParams[('hipGraphNode_t', 'node'), ('const hipKernelNodeParams*', 'pNodeParams')] case HIP_API_ID_hipGraphKernelNodeSetParams: if (data->args.hipGraphKernelNodeSetParams.pNodeParams) data->args.hipGraphKernelNodeSetParams.pNodeParams__val = *(data->args.hipGraphKernelNodeSetParams.pNodeParams); @@ -5262,6 +5325,9 @@ static inline void hipApiArgsInit(hip_api_id_t id, hip_api_data_t* data) { case HIP_API_ID_hipLaunchCooperativeKernelMultiDevice: if (data->args.hipLaunchCooperativeKernelMultiDevice.launchParamsList) data->args.hipLaunchCooperativeKernelMultiDevice.launchParamsList__val = *(data->args.hipLaunchCooperativeKernelMultiDevice.launchParamsList); break; +// hipLaunchHostFunc[('hipStream_t', 'stream'), ('hipHostFn_t', 'fn'), ('void*', 'userData')] + case HIP_API_ID_hipLaunchHostFunc: + break; // hipLaunchKernel[('const void*', 'function_address'), ('dim3', 'numBlocks'), ('dim3', 'dimBlocks'), ('void**', 'args'), ('size_t', 'sharedMemBytes'), ('hipStream_t', 'stream')] case HIP_API_ID_hipLaunchKernel: if (data->args.hipLaunchKernel.args) data->args.hipLaunchKernel.args__val = *(data->args.hipLaunchKernel.args); @@ -5709,10 +5775,6 @@ static inline void hipApiArgsInit(hip_api_id_t id, hip_api_data_t* data) { if (data->args.hipTexRefSetAddress2D.texRef) data->args.hipTexRefSetAddress2D.texRef__val = *(data->args.hipTexRefSetAddress2D.texRef); if (data->args.hipTexRefSetAddress2D.desc) data->args.hipTexRefSetAddress2D.desc__val = *(data->args.hipTexRefSetAddress2D.desc); break; -// hipTexRefSetArray[('textureReference*', 'tex'), ('hipArray_const_t', 'array'), ('unsigned int', 'flags')] - case HIP_API_ID_hipTexRefSetArray: - if (data->args.hipTexRefSetArray.tex) data->args.hipTexRefSetArray.tex__val = *(data->args.hipTexRefSetArray.tex); - break; // hipTexRefSetBorderColor[('textureReference*', 'texRef'), ('float*', 'pBorderColor')] case HIP_API_ID_hipTexRefSetBorderColor: if (data->args.hipTexRefSetBorderColor.texRef) data->args.hipTexRefSetBorderColor.texRef__val = *(data->args.hipTexRefSetBorderColor.texRef); @@ -6029,6 +6091,13 @@ static inline const char* hipApiString(hip_api_id_t id, const hip_api_data_t* da else oss << ", greatestPriority=" << data->args.hipDeviceGetStreamPriorityRange.greatestPriority__val; oss << ")"; break; + case HIP_API_ID_hipDeviceGetUuid: + oss << "hipDeviceGetUuid("; + if (data->args.hipDeviceGetUuid.uuid == NULL) oss << "uuid=NULL"; + else oss << "uuid=" << data->args.hipDeviceGetUuid.uuid__val; + oss << ", device=" << data->args.hipDeviceGetUuid.device; + oss << ")"; + break; case HIP_API_ID_hipDevicePrimaryCtxGetState: oss << "hipDevicePrimaryCtxGetState("; oss << "dev=" << data->args.hipDevicePrimaryCtxGetState.dev; @@ -6308,6 +6377,13 @@ static inline const char* hipApiString(hip_api_id_t id, const hip_api_data_t* da oss << ", deviceList=" << data->args.hipGLGetDevices.deviceList; oss << ")"; break; + case HIP_API_ID_hipGetChannelDesc: + oss << "hipGetChannelDesc("; + if (data->args.hipGetChannelDesc.desc == NULL) oss << "desc=NULL"; + else oss << "desc=" << data->args.hipGetChannelDesc.desc__val; + oss << ", array=" << data->args.hipGetChannelDesc.array; + oss << ")"; + break; case HIP_API_ID_hipGetDevice: oss << "hipGetDevice("; if (data->args.hipGetDevice.deviceId == NULL) oss << "deviceId=NULL"; @@ -6728,6 +6804,14 @@ static inline const char* hipApiString(hip_api_id_t id, const hip_api_data_t* da oss << ", flags=" << data->args.hipGraphInstantiateWithFlags.flags; oss << ")"; break; + case HIP_API_ID_hipGraphKernelNodeGetAttribute: + oss << "hipGraphKernelNodeGetAttribute("; + oss << "hNode=" << data->args.hipGraphKernelNodeGetAttribute.hNode; + oss << ", attr=" << data->args.hipGraphKernelNodeGetAttribute.attr; + if (data->args.hipGraphKernelNodeGetAttribute.value == NULL) oss << ", value=NULL"; + else oss << ", value=" << data->args.hipGraphKernelNodeGetAttribute.value__val; + oss << ")"; + break; case HIP_API_ID_hipGraphKernelNodeGetParams: oss << "hipGraphKernelNodeGetParams("; oss << "node=" << data->args.hipGraphKernelNodeGetParams.node; @@ -6735,6 +6819,14 @@ static inline const char* hipApiString(hip_api_id_t id, const hip_api_data_t* da else oss << ", pNodeParams=" << data->args.hipGraphKernelNodeGetParams.pNodeParams__val; oss << ")"; break; + case HIP_API_ID_hipGraphKernelNodeSetAttribute: + oss << "hipGraphKernelNodeSetAttribute("; + oss << "hNode=" << data->args.hipGraphKernelNodeSetAttribute.hNode; + oss << ", attr=" << data->args.hipGraphKernelNodeSetAttribute.attr; + if (data->args.hipGraphKernelNodeSetAttribute.value == NULL) oss << ", value=NULL"; + else oss << ", value=" << data->args.hipGraphKernelNodeSetAttribute.value__val; + oss << ")"; + break; case HIP_API_ID_hipGraphKernelNodeSetParams: oss << "hipGraphKernelNodeSetParams("; oss << "node=" << data->args.hipGraphKernelNodeSetParams.node; @@ -7050,6 +7142,13 @@ static inline const char* hipApiString(hip_api_id_t id, const hip_api_data_t* da oss << ", flags=" << data->args.hipLaunchCooperativeKernelMultiDevice.flags; oss << ")"; break; + case HIP_API_ID_hipLaunchHostFunc: + oss << "hipLaunchHostFunc("; + oss << "stream=" << data->args.hipLaunchHostFunc.stream; + oss << ", fn=" << data->args.hipLaunchHostFunc.fn; + oss << ", userData=" << data->args.hipLaunchHostFunc.userData; + oss << ")"; + break; case HIP_API_ID_hipLaunchKernel: oss << "hipLaunchKernel("; oss << "function_address=" << data->args.hipLaunchKernel.function_address; @@ -8039,14 +8138,6 @@ static inline const char* hipApiString(hip_api_id_t id, const hip_api_data_t* da oss << ", Pitch=" << data->args.hipTexRefSetAddress2D.Pitch; oss << ")"; break; - case HIP_API_ID_hipTexRefSetArray: - oss << "hipTexRefSetArray("; - if (data->args.hipTexRefSetArray.tex == NULL) oss << "tex=NULL"; - else oss << "tex=" << data->args.hipTexRefSetArray.tex__val; - oss << ", array=" << data->args.hipTexRefSetArray.array; - oss << ", flags=" << data->args.hipTexRefSetArray.flags; - oss << ")"; - break; case HIP_API_ID_hipTexRefSetBorderColor: oss << "hipTexRefSetBorderColor("; if (data->args.hipTexRefSetBorderColor.texRef == NULL) oss << "texRef=NULL"; diff --git a/hipamd/include/hip/nvidia_detail/nvidia_hip_runtime_api.h b/hipamd/include/hip/nvidia_detail/nvidia_hip_runtime_api.h index cb9b23aa6a..3cf47338b8 100644 --- a/hipamd/include/hip/nvidia_detail/nvidia_hip_runtime_api.h +++ b/hipamd/include/hip/nvidia_detail/nvidia_hip_runtime_api.h @@ -1,5 +1,5 @@ /* -Copyright (c) 2015 - 2021 Advanced Micro Devices, Inc. All rights reserved. +Copyright (c) 2015 - 2022 Advanced Micro Devices, Inc. All rights reserved. Permission is hereby granted, free of charge, to any person obtaining a copy of this software and associated documentation files (the "Software"), to deal @@ -2089,6 +2089,18 @@ inline static hipError_t hipDeviceGetName(char* name, int len, hipDevice_t devic return hipCUResultTohipError(cuDeviceGetName(name, len, device)); } +inline static hipError_t hipDeviceGetUuid(hipUUID* uuid, hipDevice_t device) { + if (uuid == NULL) { + return hipErrorInvalidValue; + } + struct CUuuid_st CUuid; + hipError_t err = hipCUResultTohipError(cuDeviceGetUuid(&CUuid, device)); + if (err == hipSuccess) { + strncpy(uuid->bytes, CUuid.bytes, 16); + } + return err; +} + inline static hipError_t hipDeviceGetP2PAttribute(int* value, hipDeviceP2PAttr attr, int srcDevice, int dstDevice) { return hipCUDAErrorTohipError(cudaDeviceGetP2PAttribute(value, attr, srcDevice, dstDevice)); diff --git a/hipamd/src/amdhip.def b/hipamd/src/amdhip.def index c5ed8e6a12..b7b2bb7385 100644 --- a/hipamd/src/amdhip.def +++ b/hipamd/src/amdhip.def @@ -27,6 +27,7 @@ hipDeviceGetCacheConfig hipDeviceGetStreamPriorityRange hipDeviceGetLimit hipDeviceGetName +hipDeviceGetUuid hipDeviceGetPCIBusId hipDeviceGetSharedMemConfig hipDeviceGetP2PAttribute diff --git a/hipamd/src/hip_device.cpp b/hipamd/src/hip_device.cpp index f4eab0686f..a51ca22c0c 100644 --- a/hipamd/src/hip_device.cpp +++ b/hipamd/src/hip_device.cpp @@ -149,6 +149,25 @@ hipError_t hipDeviceGetName(char *name, int len, hipDevice_t device) { HIP_RETURN(hipSuccess); } +hipError_t hipDeviceGetUuid(hipUUID* uuid, hipDevice_t device) { + HIP_INIT_API(hipDeviceGetUuid, reinterpret_cast(uuid), device); + + if (device < 0 || static_cast(device) >= g_devices.size()) { + HIP_RETURN(hipErrorInvalidDevice); + } + + if (uuid == nullptr) { + HIP_RETURN(hipErrorInvalidValue); + } + + auto* deviceHandle = g_devices[device]->devices()[0]; + const auto& info = deviceHandle->info(); + + ::strncpy(uuid->bytes, info.uuid_, 16); + + HIP_RETURN(hipSuccess); +} + hipError_t ihipGetDeviceProperties(hipDeviceProp_t* props, hipDevice_t device) { if (props == nullptr) { return hipErrorInvalidValue; diff --git a/hipamd/src/hip_hcc.def.in b/hipamd/src/hip_hcc.def.in index f3357d2418..68d6b0038c 100644 --- a/hipamd/src/hip_hcc.def.in +++ b/hipamd/src/hip_hcc.def.in @@ -27,6 +27,7 @@ hipDeviceGetCacheConfig hipDeviceGetStreamPriorityRange hipDeviceGetLimit hipDeviceGetName +hipDeviceGetUuid hipDeviceGetPCIBusId hipDeviceGetSharedMemConfig hipDeviceGetP2PAttribute diff --git a/hipamd/src/hip_hcc.map.in b/hipamd/src/hip_hcc.map.in index d2986095cf..982fab1445 100644 --- a/hipamd/src/hip_hcc.map.in +++ b/hipamd/src/hip_hcc.map.in @@ -396,3 +396,10 @@ global: local: *; } hip_4.5; + +hip_5.1 { +global: + hipDeviceGetUuid; +local: + *; +} hip_5.0;