updating roctx documentation for functions (#30)
updating roctx documentation for funcitons
Этот коммит содержится в:
коммит произвёл
GitHub
родитель
b7661bccfd
Коммит
6d2e70d8da
@@ -276,6 +276,14 @@ Here is a list of useful APIs for code instrumentation.
|
||||
- ``roctxRangePush``: Starts a new nested range.
|
||||
- ``roctxRangePop``: Stops the current nested range.
|
||||
- ``roctxRangeStop``: Stops the given range.
|
||||
- ``roctxProfilerPause``: Request any currently running profiling tool that it should stop collecting data.
|
||||
- ``roctxProfilerResume``: Request any currently running profiling tool that it should resume collecting data.
|
||||
- ``roctxGetThreadId``: Retrieve a id value for the current thread which will be identical to the id value a profiling tool gets via `rocprofiler_get_thread_id(rocprofiler_thread_id_t*)`.
|
||||
- ``roctxNameOsThread``: Current CPU OS thread to be labeled by the provided name in the output of the profiling tool.
|
||||
- ``roctxNameHsaAgent``: Given HSA agent to be labeled by the provided name in the output of the profiling tool.
|
||||
- ``roctxNameHipDevice``: Given HIP device id to be labeled by the provided name in the output of the profiling tool.
|
||||
- ``roctxNameHipStream``: Given HIP stream to be labeled by the provided name in the output of the profiling tool.
|
||||
|
||||
|
||||
.. note::
|
||||
To use ``rocprofv3`` for marker tracing, including and linking to old ROCTx works but it is recommended to switch to new ROCTx because
|
||||
@@ -328,6 +336,59 @@ Here are the contents of ``marker_api_trace.csv`` file:
|
||||
|
||||
For the description of the fields in the output file, see :ref:`output-file-fields`.
|
||||
|
||||
``roctxProfilerPause`` and ``roctxProfilerResume`` can be used to hide the calls between them. This is useful when you want to hide the calls that are not relevant to your profiling session.
|
||||
|
||||
.. code-block:: bash
|
||||
|
||||
#include <rocprofiler-sdk-roctx/roctx.h>
|
||||
|
||||
// Memory transfer from host to device
|
||||
HIP_API_CALL(hipMemcpy(gpuMatrix, Matrix, NUM * sizeof(float), hipMemcpyHostToDevice));
|
||||
|
||||
auto tid = roctx_thread_id_t{};
|
||||
roctxGetThreadId(&tid);
|
||||
roctxProfilerPause(tid);
|
||||
// Memory transfer that should be hidden by profiling tool
|
||||
HIP_API_CALL(
|
||||
hipMemcpy(gpuTransposeMatrix, gpuMatrix, NUM * sizeof(float), hipMemcpyDeviceToDevice));
|
||||
roctxProfilerResume(tid);
|
||||
|
||||
// Lauching kernel from host
|
||||
hipLaunchKernelGGL(matrixTranspose,
|
||||
dim3(WIDTH / THREADS_PER_BLOCK_X, WIDTH / THREADS_PER_BLOCK_Y),
|
||||
dim3(THREADS_PER_BLOCK_X, THREADS_PER_BLOCK_Y),
|
||||
0,
|
||||
0,
|
||||
gpuTransposeMatrix,
|
||||
gpuMatrix,
|
||||
WIDTH);
|
||||
|
||||
// Memory transfer from device to host
|
||||
HIP_API_CALL(
|
||||
hipMemcpy(TransposeMatrix, gpuTransposeMatrix, NUM * sizeof(float), hipMemcpyDeviceToHost));
|
||||
|
||||
.. code-block:: shell
|
||||
|
||||
rocprofv3 --marker-trace --hip-trace -- <application_path>
|
||||
|
||||
The above command generates a ``hip_api_trace.csv`` file prefixed with the process ID, which has only 2 `hipMemcpy` calls and the in between ``hipMemcpyDeviceToHost`` is hidden .
|
||||
|
||||
.. code-block:: shell
|
||||
|
||||
"Domain","Function","Process_Id","Thread_Id","Correlation_Id","Start_Timestamp","End_Timestamp"
|
||||
"HIP_COMPILER_API","__hipRegisterFatBinary",1643920,1643920,1,320301257609216,320301257636427
|
||||
"HIP_COMPILER_API","__hipRegisterFunction",1643920,1643920,2,320301257650707,320301257678857
|
||||
"HIP_RUNTIME_API","hipGetDevicePropertiesR0600",1643920,1643920,4,320301258114239,320301337764472
|
||||
"HIP_RUNTIME_API","hipMalloc",1643920,1643920,5,320301338073823,320301338247374
|
||||
"HIP_RUNTIME_API","hipMalloc",1643920,1643920,6,320301338248284,320301338399595
|
||||
"HIP_RUNTIME_API","hipMemcpy",1643920,1643920,7,320301338410995,320301631549262
|
||||
"HIP_COMPILER_API","__hipPushCallConfiguration",1643920,1643920,10,320301632131175,320301632134215
|
||||
"HIP_COMPILER_API","__hipPopCallConfiguration",1643920,1643920,11,320301632137745,320301632139735
|
||||
"HIP_RUNTIME_API","hipLaunchKernel",1643920,1643920,12,320301632142615,320301632898289
|
||||
"HIP_RUNTIME_API","hipMemcpy",1643920,1643920,14,320301632901249,320301633934395
|
||||
"HIP_RUNTIME_API","hipFree",1643920,1643920,15,320301643320908,320301643511479
|
||||
"HIP_RUNTIME_API","hipFree",1643920,1643920,16,320301643512629,320301643585639
|
||||
|
||||
Kernel Rename
|
||||
++++++++++++++
|
||||
|
||||
|
||||
Ссылка в новой задаче
Block a user