AI/rocm-systems
2
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

SWDEV-323472 - Add hipThreadExchangeStreamCaptureMode

Browse Source 
Also fix an issue in hip_prof_gen.py with GLenum.

Change-Id: Ib03fe9dae87b850c53b7d8aeb78b8dbdd01f794c


[ROCm/clr commit: 27f102b167]
This commit is contained in:
Christophe Paquot
2022-02-18 16:12:12 -08:00
committed by Saleel Kudchadker
parent 57b1cd0cf2
commit 7052f1f5de
6 changed files with 50 additions and 59 deletions
Download Patch File Download Diff File Expand all files Collapse all files
projects/clr/hipamd/src/amdhip.def
+1
View File
@@ -354,3 +354,4 @@ hipGraphExecMemsetNodeSetParams
amd_dbgapi_get_build_name
amd_dbgapi_get_git_hash
amd_dbgapi_get_build_id
hipThreadExchangeStreamCaptureMode
projects/clr/hipamd/src/hip_graph.cpp
+17
View File
@@ -695,6 +695,23 @@ hipError_t hipStreamIsCapturing(hipStream_t stream, hipStreamCaptureStatus* pCap
HIP_RETURN(hipSuccess);
}
hipError_t hipThreadExchangeStreamCaptureMode(hipStreamCaptureMode* mode) {
HIP_INIT_API(hipThreadExchangeStreamCaptureMode, mode);
if (mode == nullptr ||
*mode < hipStreamCaptureModeGlobal ||
*mode > hipStreamCaptureModeRelaxed ||
g_captureStreams.size() == 0) {
HIP_RETURN(hipErrorInvalidValue);
}
hipStreamCaptureMode oldMode = reinterpret_cast<hip::Stream*>(g_captureStreams[0])->GetCaptureMode();
reinterpret_cast<hip::Stream*>(g_captureStreams[0])->SetCaptureMode(*mode);
*mode = oldMode;
HIP_RETURN_DURATION(hipSuccess);
}
hipError_t hipStreamBeginCapture(hipStream_t stream, hipStreamCaptureMode mode) {
HIP_INIT_API(hipStreamBeginCapture, stream, mode);
if (!hip::isValid(stream)) {
projects/clr/hipamd/src/hip_hcc.def.in
+1
View File
@@ -352,3 +352,4 @@ amd_dbgapi_get_git_hash
amd_dbgapi_get_build_id
hipStreamGetCaptureInfo
hipStreamGetCaptureInfo_v2
hipThreadExchangeStreamCaptureMode
projects/clr/hipamd/src/hip_hcc.map.in
+1
View File
@@ -392,6 +392,7 @@ hip_5.0 {
global:
hipPointerGetAttribute;
hipDrvPointerGetAttributes;
hipThreadExchangeStreamCaptureMode;
local:
*;
} hip_4.5;
projects/clr/hipamd/src/hip_prof_gen.py
+5 -3
View File
@@ -81,7 +81,7 @@ def filtr_api_args(args_str):
args_str = re.sub(r'\s*,\s*', r',', args_str);
args_str = re.sub(r'\s+', r' ', args_str);
args_str = re.sub(r'\s*(\*+)\s*', r'\1 ', args_str);
args_str = re.sub(r'(enum|struct) ', '', args_str);
args_str = re.sub(r'(\benum|struct) ', '', args_str);
return args_str
# Normalizing types
@@ -129,7 +129,9 @@ def pointer_ck(arg_type):
m = re.match(r'(.*)\*$', arg_type)
if m:
ptr_type = m.group(1)
ptr_type = re.sub(r'const ', '', ptr_type)
n = re.match(r'(.*)\*\*$', arg_type)
if not n:
ptr_type = re.sub(r'const ', '', ptr_type)
if ptr_type == 'void': ptr_type = ''
return ptr_type
#############################################################
@@ -613,7 +615,7 @@ if (len(sys.argv) < 4):
" -p - HIP_INIT_API macro patching mode\n" +
"\n" +
" Example:\n" +
" $ " + sys.argv[0] + " -v -p -t --priv ./include/hip/amd_detail/hip_runtime_api.h" +
" $ " + sys.argv[0] + " -v -p -t --priv ../hip/include/hip/hip_runtime_api.h" +
" ./src ./include/hip/amd_detail/hip_prof_str.h ./include/hip/amd_detail/hip_prof_str.h.new");
# API header file given as an argument