From eb769499b63474f70c4dcf752fcfde92867d13eb Mon Sep 17 00:00:00 2001 From: Ammar ELWazir Date: Tue, 20 Jun 2023 20:51:47 +0000 Subject: [PATCH] SWDEV-405920: V2 clean up for unimplemented and corrupted APIs Change-Id: I2c88d8415a3a386377e99b93295adb82e1f7195e [ROCm/rocprofiler commit: 37c48ce0ab02b2d6eb1c03c98b8e70decacef524] --- projects/rocprofiler/CHANGELOG.md | 1 + .../include/rocprofiler/v2/rocprofiler.h | 84 ------------------- projects/rocprofiler/samples/CMakeLists.txt | 27 ------ .../pcie_counters_example.cpp | 6 +- .../xgmi_counters_sampler_example.cpp | 2 +- .../profiler/application_replay_sample.cpp | 72 ---------------- .../samples/profiler/kernel_replay_sample.cpp | 68 --------------- .../samples/profiler/user_replay_sample.cpp | 78 ----------------- .../rocprofiler/samples/tracer/sample.cpp | 2 +- .../samples/tracer/sample_async.cpp | 2 +- projects/rocprofiler/src/api/exportmap | 4 - .../rocprofiler/src/api/rocprofilerv2.cpp | 42 ---------- projects/rocprofiler/src/tools/tool.cpp | 2 +- .../featuretests/profiler/profiler_gtest.cpp | 6 +- .../tests-v2/unittests/profiler_gtest.cpp | 23 ----- 15 files changed, 11 insertions(+), 408 deletions(-) delete mode 100644 projects/rocprofiler/samples/profiler/application_replay_sample.cpp delete mode 100644 projects/rocprofiler/samples/profiler/kernel_replay_sample.cpp delete mode 100644 projects/rocprofiler/samples/profiler/user_replay_sample.cpp diff --git a/projects/rocprofiler/CHANGELOG.md b/projects/rocprofiler/CHANGELOG.md index 7739e350b3..25a0c07965 100644 --- a/projects/rocprofiler/CHANGELOG.md +++ b/projects/rocprofiler/CHANGELOG.md @@ -226,6 +226,7 @@ Rocprofiler for ROCm 5.7 added support for counter collection (PMC) and advanced ### Changed - ATT analysis will not run by default. For ATT to have the same behaviour as 5.5, use --plugin att --mode network - Kernel Names are now removed from HIP API records, users of the API can get the kernel names from the corresponding HIP Dispatch OPS using the correlation ID, this change was done to optimize and to manage the data copied. +- Removing Replay modes as we discovered that some of them will corrupt the applications' behavior, we will re-add them once we implement the fix for them. ### Optimized - ATT json filesizes diff --git a/projects/rocprofiler/include/rocprofiler/v2/rocprofiler.h b/projects/rocprofiler/include/rocprofiler/v2/rocprofiler.h index b32835aefd..09b7a22952 100644 --- a/projects/rocprofiler/include/rocprofiler/v2/rocprofiler.h +++ b/projects/rocprofiler/include/rocprofiler/v2/rocprofiler.h @@ -1568,7 +1568,6 @@ ROCPROFILER_API rocprofiler_status_t rocprofiler_next_record( * @{ */ -// TODO(aelwazir): Replay mode naming !!?? (If changed, reflect on start&stop) /** * Replay Profiling Modes. */ @@ -1578,18 +1577,6 @@ typedef enum { * sure that no replays will be done */ ROCPROFILER_NONE_REPLAY_MODE = -1, - /** - * Replaying the whole application to get multi passes (Not Yet Supported) - */ - ROCPROFILER_APPLICATION_REPLAY_MODE = 0, - /** - * Replaying every kernel dispatch to get multi passes - */ - ROCPROFILER_KERNEL_REPLAY_MODE = 1, - /** - * Replaying an user-specified range to get multi passes (Not Yet Supported) - */ - ROCPROFILER_USER_REPLAY_MODE = 2 } rocprofiler_replay_mode_t; /** @@ -2149,77 +2136,6 @@ ROCPROFILER_API rocprofiler_status_t rocprofiler_start_session(rocprofiler_sessi ROCPROFILER_API rocprofiler_status_t rocprofiler_terminate_session(rocprofiler_session_id_t session_id) ROCPROFILER_VERSION_9_0; -/** \defgroup session_range_group Session Range Labeling - * \ingroup sessions_handling_group - * @{ - */ - -/** - * Setting a label to a block range - * This can be used to label a range of code that is having active profiling - * session or labeling a pass - * - * \param[in] label The label given for a certain block or pass to name/label. - * \retval ::ROCPROFILER_STATUS_SUCCESS The function has been executed - * successfully. - * \retval ::ROCPROFILER_STATUS_ERROR_NOT_INITIALIZED, if rocprofiler_initialize - * wasn't called before or if rocprofiler_finalize is called - * \retval ::ROCPROFILER_STATUS_ERROR_CORRUPTED_LABEL_DATA \n may return if - * the label pointer can't be read by the API - */ -ROCPROFILER_API rocprofiler_status_t rocprofiler_push_range(const char* label) - ROCPROFILER_VERSION_9_0; - -/** - * Setting an endpoint for a range - * This function can be used to set an endpoint to range labeled by - * ::rocprofiler_push_range - * - * \retval ::ROCPROFILER_STATUS_SUCCESS The function has been executed - * successfully - * \retval ::ROCPROFILER_STATUS_ERROR_NOT_INITIALIZED, if rocprofiler_initialize - * wasn't called before or if rocprofiler_finalize is called - * \retval ::ROCPROFILER_STATUS_ERROR_RANGE_STACK_IS_EMPTY may return if - * ::rocprofiler_push_range wasn't called correctly - */ -ROCPROFILER_API rocprofiler_status_t rocprofiler_pop_range() ROCPROFILER_VERSION_9_0; - -/** @} */ - -/** \defgroup session_user_replay_pass_group Session User Replay Pass Mode - * \ingroup sessions_handling_group - * @{ - */ - -/** - * Create and Start a pass - * A Pass is a block of code that can be replayed if required by the - * profiling/tracing and it mainly depends on the profiling data given in the - * ::rocprofiler_create_session - * - * \retval ::ROCPROFILER_STATUS_SUCCESS The function has been executed - * successfully. - * \retval ::ROCPROFILER_STATUS_ERROR_NOT_INITIALIZED, if rocprofiler_initialize - * wasn't called before or if rocprofiler_finalize is called - * \retval ::ROCPROFILER_STATUS_ERROR_SESSION_NOT_FOUND \n If the no active session - * found - */ -ROCPROFILER_API rocprofiler_status_t rocprofiler_start_replay_pass() ROCPROFILER_VERSION_9_0; - -/** - * End a pass - * End a pass created and started by ::rocprofiler_start_pass - * - * \retval ::ROCPROFILER_STATUS_SUCCESS The function has been executed - * successfully. - * \retval ::ROCPROFILER_STATUS_ERROR_NOT_INITIALIZED, if rocprofiler_initialize - * wasn't called before or if rocprofiler_finalize is called - * \retval ::ROCPROFILER_STATUS_ERROR_PASS_NOT_STARTED \n if there is no pass - * started before this call - */ -ROCPROFILER_API rocprofiler_status_t rocprofiler_end_replay_pass() ROCPROFILER_VERSION_9_0; - -/** @} */ /** @} */ /** \defgroup device_profiling Device Profiling API diff --git a/projects/rocprofiler/samples/CMakeLists.txt b/projects/rocprofiler/samples/CMakeLists.txt index 8d2f918207..28369f23bb 100644 --- a/projects/rocprofiler/samples/CMakeLists.txt +++ b/projects/rocprofiler/samples/CMakeLists.txt @@ -48,24 +48,6 @@ file(GLOB ROCPROFILER_UTIL_SRC_FILES ${PROJECT_SOURCE_DIR}/src/utils/helper.cpp) # Profiler Samples # ############################################################################################################################################ -## Build Application Replay Sample -set_source_files_properties(profiler/application_replay_sample.cpp PROPERTIES HIP_SOURCE_PROPERTY_FORMAT 1) -hip_add_executable(profiler_application_replay profiler/application_replay_sample.cpp ${ROCPROFILER_UTIL_SRC_FILES}) -target_include_directories(profiler_application_replay PRIVATE ${PROJECT_SOURCE_DIR} ${CMAKE_CURRENT_SOURCE_DIR}/common) -target_link_libraries(profiler_application_replay PRIVATE rocprofiler-v2 amd_comgr) -target_link_options(profiler_application_replay PRIVATE "-Wl,--build-id=md5") -add_dependencies(samples profiler_application_replay) -install(TARGETS profiler_application_replay RUNTIME DESTINATION ${CMAKE_INSTALL_DATAROOTDIR}/${PROJECT_NAME}/samples COMPONENT samples) - -## Build Kernel Replay Sample -set_source_files_properties(profiler/kernel_replay_sample.cpp PROPERTIES HIP_SOURCE_PROPERTY_FORMAT 1) -hip_add_executable(profiler_kernel_replay profiler/kernel_replay_sample.cpp ${ROCPROFILER_UTIL_SRC_FILES}) -target_include_directories(profiler_kernel_replay PRIVATE ${PROJECT_SOURCE_DIR} ${CMAKE_CURRENT_SOURCE_DIR}/common) -target_link_libraries(profiler_kernel_replay PRIVATE rocprofiler-v2 amd_comgr) -target_link_options(profiler_kernel_replay PRIVATE "-Wl,--build-id=md5") -add_dependencies(samples profiler_kernel_replay) -install(TARGETS profiler_kernel_replay RUNTIME DESTINATION ${CMAKE_INSTALL_DATAROOTDIR}/${PROJECT_NAME}/samples COMPONENT samples) - ## Build Kernel No Replay Sample set_source_files_properties(profiler/kernel_profiling_no_replay_sample.cpp PROPERTIES HIP_SOURCE_PROPERTY_FORMAT 1) hip_add_executable(profiler_kernel_no_replay profiler/kernel_profiling_no_replay_sample.cpp ${ROCPROFILER_UTIL_SRC_FILES}) @@ -75,15 +57,6 @@ target_link_options(profiler_kernel_no_replay PRIVATE "-Wl,--build-id=md5") add_dependencies(samples profiler_kernel_no_replay) install(TARGETS profiler_kernel_no_replay RUNTIME DESTINATION ${CMAKE_INSTALL_DATAROOTDIR}/${PROJECT_NAME}/samples COMPONENT samples) -## Build User Replay Sample -set_source_files_properties(profiler/user_replay_sample.cpp PROPERTIES HIP_SOURCE_PROPERTY_FORMAT 1) -hip_add_executable(profiler_user_replay profiler/user_replay_sample.cpp ${ROCPROFILER_UTIL_SRC_FILES}) -target_include_directories(profiler_user_replay PRIVATE ${PROJECT_SOURCE_DIR} ${CMAKE_CURRENT_SOURCE_DIR}/common) -target_link_options(profiler_user_replay PRIVATE "-Wl,--build-id=md5") -target_link_libraries(profiler_user_replay PRIVATE rocprofiler-v2 amd_comgr) -add_dependencies(samples profiler_user_replay) -install(TARGETS profiler_user_replay RUNTIME DESTINATION ${CMAKE_INSTALL_DATAROOTDIR}/${PROJECT_NAME}/samples COMPONENT samples) - ## Build Device Profiling Sample set_source_files_properties(profiler/device_profiling_sample.cpp PROPERTIES HIP_SOURCE_PROPERTY_FORMAT 1) hip_add_executable(profiler_device_profiling profiler/device_profiling_sample.cpp ${ROCPROFILER_UTIL_SRC_FILES}) diff --git a/projects/rocprofiler/samples/counters_sampler/pcie_counters_example.cpp b/projects/rocprofiler/samples/counters_sampler/pcie_counters_example.cpp index da070d277c..80dbae8670 100644 --- a/projects/rocprofiler/samples/counters_sampler/pcie_counters_example.cpp +++ b/projects/rocprofiler/samples/counters_sampler/pcie_counters_example.cpp @@ -3,7 +3,7 @@ int main(int argc, char** argv) { int* gpuMem; int counter_option = 0; - + std::vector pcie_counters = { "CI_PERF_slv_MemRd_Bandwidth0", "CI_PERF_slv_MemWr_Bandwidth0", "CI_PERF_slv_totalMemRdTx", "CI_PERF_slv_totalMemWrTx", "CI_PERF_slv_totalTx"}; @@ -26,7 +26,7 @@ int main(int argc, char** argv) { // Creating the session with given replay mode rocprofiler_session_id_t session_id; - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); + CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id)); // Creating Output Buffer for the data rocprofiler_buffer_id_t buffer_id; @@ -47,7 +47,7 @@ int main(int argc, char** argv) { {.name = const_cast(pcie_counters[counter_option].c_str()), .type = ROCPROFILER_COUNTERS_SAMPLER_PCIE_COUNTERS}}; - uint32_t rate = 1000; + uint32_t rate = 1000; uint32_t duration = 5000; rocprofiler_counters_sampler_parameters_t cs_parameters = {.counters = counters_input, diff --git a/projects/rocprofiler/samples/counters_sampler/xgmi_counters_sampler_example.cpp b/projects/rocprofiler/samples/counters_sampler/xgmi_counters_sampler_example.cpp index e360fd92a9..360245c5ed 100644 --- a/projects/rocprofiler/samples/counters_sampler/xgmi_counters_sampler_example.cpp +++ b/projects/rocprofiler/samples/counters_sampler/xgmi_counters_sampler_example.cpp @@ -11,7 +11,7 @@ int main(int argc, char** argv) { // Creating the session with given replay mode rocprofiler_session_id_t session_id; - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); + CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id)); // Creating Output Buffer for the data rocprofiler_buffer_id_t buffer_id; diff --git a/projects/rocprofiler/samples/profiler/application_replay_sample.cpp b/projects/rocprofiler/samples/profiler/application_replay_sample.cpp deleted file mode 100644 index 9185a6fb0b..0000000000 --- a/projects/rocprofiler/samples/profiler/application_replay_sample.cpp +++ /dev/null @@ -1,72 +0,0 @@ -#include "../common/common.h" - -int main(int argc, char** argv) { - printf("APPLICATION REPLAY Mode is not yet Supported!"); -#if 0 - int* gpuMem; - prepare(); - // Initialize the tools - CHECK_ROCPROFILER(rocprofiler_initialize()); - - // Creating the session with given replay mode - rocprofiler_session_id_t session_id; - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_APPLICATION_REPLAY_MODE, &session_id)); - - // Creating Output Buffer for the data - rocprofiler_buffer_id_t buffer_id; - CHECK_ROCPROFILER(rocprofiler_create_buffer( - session_id, - [](const rocprofiler_record_header_t* record, const rocprofiler_record_header_t* end_record, - rocprofiler_session_id_t session_id, rocprofiler_buffer_id_t buffer_id) { - WriteBufferRecords(record, end_record, session_id, buffer_id); - }, - 0x9999, &buffer_id)); - - // Counter Collection Filter - std::vector counters; - counters.emplace_back("GRBM_COUNT"); - rocprofiler_filter_id_t filter_id; - [[maybe_unused]] rocprofiler_filter_property_t property = {}; - CHECK_ROCPROFILER(rocprofiler_create_filter(session_id, ROCPROFILER_COUNTERS_COLLECTION, - rocprofiler_filter_data_t{.counters_names = &counters[0]}, - counters.size(), &filter_id, property)); - CHECK_ROCPROFILER(rocprofiler_set_filter_buffer(session_id, filter_id, buffer_id)); - filter_ids.emplace_back(filter_id); - - // Normal HIP Calls - hipDeviceProp_t devProp; - HIP_CALL(hipGetDeviceProperties(&devProp, 0)); - HIP_CALL(hipMalloc((void**)&gpuMem, 1 * sizeof(int))); - - // KernelA and KernelB won't be profiled - kernelCalls('A'); - kernelCalls('B'); - - // Activating Profiling Session to profile whatever kernel launches occurs up - // till the next terminate session - CHECK_ROCPROFILER(rocprofiler_start_session(session_id)); - - // KernelC, KernelD, KernelE and KernelF to be profiled as part of the session - kernelCalls('C'); - kernelCalls('D'); - kernelCalls('E'); - kernelCalls('F'); - // Normal HIP Calls - HIP_CALL(hipFree(gpuMem)); - - // Deactivating session - CHECK_ROCPROFILER(rocprofiler_terminate_session(session_id)); - - // Manual Flush user buffer request - CHECK_ROCPROFILER(rocprofiler_flush_data(session_id, buffer_id)); - - // Destroy sessions - CHECK_ROCPROFILER(rocprofiler_destroy_session(session_id)); - - // Destroy all profiling related objects(User buffer, sessions, filters, - // etc..) - CHECK_ROCPROFILER(rocprofiler_finalize()); -#endif - - return 0; -} \ No newline at end of file diff --git a/projects/rocprofiler/samples/profiler/kernel_replay_sample.cpp b/projects/rocprofiler/samples/profiler/kernel_replay_sample.cpp deleted file mode 100644 index 492100163e..0000000000 --- a/projects/rocprofiler/samples/profiler/kernel_replay_sample.cpp +++ /dev/null @@ -1,68 +0,0 @@ -#include "../common/common.h" - -int main(int argc, char** argv) { - int* gpuMem; - prepare(); - // Initialize the tools - CHECK_ROCPROFILER(rocprofiler_initialize()); - - // Creating the session with given replay mode - rocprofiler_session_id_t session_id; - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); - - // Creating Output Buffer for the data - rocprofiler_buffer_id_t buffer_id; - CHECK_ROCPROFILER(rocprofiler_create_buffer( - session_id, - [](const rocprofiler_record_header_t* record, const rocprofiler_record_header_t* end_record, - rocprofiler_session_id_t session_id, rocprofiler_buffer_id_t buffer_id) { - WriteBufferRecords(record, end_record, session_id, buffer_id); - }, - 0x9999, &buffer_id)); - - // Counter Collection Filter - std::vector counters; - counters.emplace_back("GRBM_COUNT"); - rocprofiler_filter_id_t filter_id; - [[maybe_unused]] rocprofiler_filter_property_t property = {}; - CHECK_ROCPROFILER(rocprofiler_create_filter(session_id, ROCPROFILER_COUNTERS_COLLECTION, - rocprofiler_filter_data_t{.counters_names = &counters[0]}, - counters.size(), &filter_id, property)); - CHECK_ROCPROFILER(rocprofiler_set_filter_buffer(session_id, filter_id, buffer_id)); - - // Normal HIP Calls - hipDeviceProp_t devProp; - HIP_CALL(hipGetDeviceProperties(&devProp, 0)); - HIP_CALL(hipMalloc((void**)&gpuMem, 1 * sizeof(int))); - - // KernelA and KernelB won't be profiled - kernelCalls('A'); - kernelCalls('B'); - - // Activating Profiling Session to profile whatever kernel launches occurs up - // till the next terminate session - CHECK_ROCPROFILER(rocprofiler_start_session(session_id)); - - // KernelC, KernelD, KernelE and KernelF to be profiled as part of the session - kernelCalls('C'); - kernelCalls('D'); - kernelCalls('E'); - kernelCalls('F'); - // Normal HIP Calls - HIP_CALL(hipFree(gpuMem)); - - // Deactivating session - CHECK_ROCPROFILER(rocprofiler_terminate_session(session_id)); - - // Manual Flush user buffer request - CHECK_ROCPROFILER(rocprofiler_flush_data(session_id, buffer_id)); - - // Destroy sessions - CHECK_ROCPROFILER(rocprofiler_destroy_session(session_id)); - - // Destroy all profiling related objects(User buffer, sessions, filters, - // etc..) - CHECK_ROCPROFILER(rocprofiler_finalize()); - - return 0; -} \ No newline at end of file diff --git a/projects/rocprofiler/samples/profiler/user_replay_sample.cpp b/projects/rocprofiler/samples/profiler/user_replay_sample.cpp deleted file mode 100644 index d20d37c6fe..0000000000 --- a/projects/rocprofiler/samples/profiler/user_replay_sample.cpp +++ /dev/null @@ -1,78 +0,0 @@ -#include "../common/common.h" - -int main(int argc, char** argv) { - printf("USER REPLAY Mode is not yet Supported!"); -#if 0 - int* gpuMem; - prepare(); - // Initialize the tools - CHECK_ROCPROFILER(rocprofiler_initialize()); - - // Creating the session with given replay mode - rocprofiler_session_id_t session_id; - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_USER_REPLAY_MODE, &session_id)); - - // Creating Output Buffer for the data - rocprofiler_buffer_id_t buffer_id; - CHECK_ROCPROFILER(rocprofiler_create_buffer( - session_id, - [](const rocprofiler_record_header_t* record, const rocprofiler_record_header_t* end_record, - rocprofiler_session_id_t session_id, rocprofiler_buffer_id_t buffer_id) { - WriteBufferRecords(record, end_record, session_id, buffer_id); - }, - 0x9999, &buffer_id)); - - // Counter Collection Filter - std::vector counters; - counters.emplace_back("GRBM_COUNT"); - rocprofiler_filter_id_t filter_id; - [[maybe_unused]] rocprofiler_filter_property_t property = {}; - CHECK_ROCPROFILER(rocprofiler_create_filter(session_id, ROCPROFILER_COUNTERS_COLLECTION, - rocprofiler_filter_data_t{.counters_names = &counters[0]}, - counters.size(), &filter_id, property)); - CHECK_ROCPROFILER(rocprofiler_set_filter_buffer(session_id, filter_id, buffer_id)); - filter_ids.emplace_back(filter_id); - - // Normal HIP Calls - hipDeviceProp_t devProp; - HIP_CALL(hipGetDeviceProperties(&devProp, 0)); - HIP_CALL(hipMalloc((void**)&gpuMem, 1 * sizeof(int))); - - // KernelA and KernelB won't be profiled - kernelCalls('A'); - kernelCalls('B'); - - // Activating Profiling Session to profile whatever kernel launches occurs up - // till the next terminate session - CHECK_ROCPROFILER(rocprofiler_start_session(session_id)); - - // Replay Pass Start point - CHECK_ROCPROFILER(rocprofiler_start_replay_pass()); - - // KernelC, KernelD, KernelE and KernelF to be profiled as part of the session - kernelCalls('C'); - kernelCalls('D'); - kernelCalls('E'); - kernelCalls('F'); - // Normal HIP Calls - HIP_CALL(hipFree(gpuMem)); - - // Deactivating session - CHECK_ROCPROFILER(rocprofiler_terminate_session(session_id)); - - // Manual Flush user buffer request - CHECK_ROCPROFILER(rocprofiler_flush_data(session_id, buffer_id)); - - // Replay Pass End point - CHECK_ROCPROFILER(rocprofiler_end_replay_pass()); - - // Destroy sessions - CHECK_ROCPROFILER(rocprofiler_destroy_session(session_id)); - - // Destroy all profiling related objects(User buffer, sessions, filters, - // etc..) - CHECK_ROCPROFILER(rocprofiler_finalize()); -#endif - - return 0; -} diff --git a/projects/rocprofiler/samples/tracer/sample.cpp b/projects/rocprofiler/samples/tracer/sample.cpp index 02d2afe342..1a34491e4b 100644 --- a/projects/rocprofiler/samples/tracer/sample.cpp +++ b/projects/rocprofiler/samples/tracer/sample.cpp @@ -8,7 +8,7 @@ int main(int argc, char** argv) { // Creating the session with given replay mode rocprofiler_session_id_t session_id; - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); + CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id)); // Creating Output Buffer for the data rocprofiler_buffer_id_t buffer_id; diff --git a/projects/rocprofiler/samples/tracer/sample_async.cpp b/projects/rocprofiler/samples/tracer/sample_async.cpp index d769395cb8..fe57bcd018 100644 --- a/projects/rocprofiler/samples/tracer/sample_async.cpp +++ b/projects/rocprofiler/samples/tracer/sample_async.cpp @@ -8,7 +8,7 @@ int main(int argc, char** argv) { // Creating the session with given replay mode rocprofiler_session_id_t session_id; - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); + CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id)); // Creating Output Buffer for the data rocprofiler_buffer_id_t buffer_id; diff --git a/projects/rocprofiler/src/api/exportmap b/projects/rocprofiler/src/api/exportmap index d5b77ade42..0977a8d25f 100644 --- a/projects/rocprofiler/src/api/exportmap +++ b/projects/rocprofiler/src/api/exportmap @@ -33,10 +33,6 @@ global: OnLoad; rocprofiler_set_buffer_properties; rocprofiler_start_session; rocprofiler_terminate_session; - rocprofiler_push_range; - rocprofiler_pop_range; - rocprofiler_start_replay_pass; - rocprofiler_end_replay_pass; rocprofiler_device_profiling_session_create; rocprofiler_device_profiling_session_start; rocprofiler_device_profiling_session_poll; diff --git a/projects/rocprofiler/src/api/rocprofilerv2.cpp b/projects/rocprofiler/src/api/rocprofilerv2.cpp index a0de9742a2..df3f3d8102 100644 --- a/projects/rocprofiler/src/api/rocprofilerv2.cpp +++ b/projects/rocprofiler/src/api/rocprofilerv2.cpp @@ -502,48 +502,6 @@ rocprofiler_terminate_session(rocprofiler_session_id_t session_id) { API_METHOD_SUFFIX } - -// API to push a custom label for defining a code section -ROCPROFILER_API rocprofiler_status_t rocprofiler_push_range(rocprofiler_session_id_t session_id, - const char* label) { - API_INIT_CHECKER - if (!label) throw rocprofiler::Exception(ROCPROFILER_STATUS_ERROR_CORRUPTED_LABEL_DATA); - rocprofiler::GetROCProfilerSingleton()->GetSession(session_id)->PushRangeLabels(label); - API_METHOD_SUFFIX -} - -// API to pop a custom label defined for a code section -ROCPROFILER_API rocprofiler_status_t rocprofiler_pop_range(rocprofiler_session_id_t session_id) { - API_INIT_CHECKER - if (!rocprofiler::GetROCProfilerSingleton()->GetSession(session_id)->PopRangeLabels()) - throw rocprofiler::Exception(ROCPROFILER_STATUS_ERROR_RANGE_STACK_IS_EMPTY); - API_METHOD_SUFFIX -} - -ROCPROFILER_API rocprofiler_status_t -rocprofiler_start_replay_pass(rocprofiler_session_id_t session_id) { - API_INIT_CHECKER - if (!rocprofiler::GetROCProfilerSingleton()->FindSession(session_id)) - throw rocprofiler::Exception(ROCPROFILER_STATUS_ERROR_SESSION_NOT_FOUND); - rocprofiler::GetROCProfilerSingleton() - ->GetSession(session_id) - ->GetProfiler() - ->StartReplayPass(session_id); - API_METHOD_SUFFIX -} - -ROCPROFILER_API rocprofiler_status_t -rocprofiler_end_replay_pass(rocprofiler_session_id_t session_id) { - API_INIT_CHECKER - if (!rocprofiler::GetROCProfilerSingleton() - ->GetSession(session_id) - ->GetProfiler() - ->HasActivePass()) - throw rocprofiler::Exception(ROCPROFILER_STATUS_ERROR_PASS_NOT_STARTED); - rocprofiler::GetROCProfilerSingleton()->GetSession(session_id)->GetProfiler()->EndReplayPass(); - API_METHOD_SUFFIX -} - ROCPROFILER_API rocprofiler_status_t rocprofiler_device_profiling_session_create( const char** counter_names, uint64_t num_counters, rocprofiler_session_id_t* session_id, int cpu_index, int gpu_index) { diff --git a/projects/rocprofiler/src/tools/tool.cpp b/projects/rocprofiler/src/tools/tool.cpp index b28b76c42e..c8f343431d 100644 --- a/projects/rocprofiler/src/tools/tool.cpp +++ b/projects/rocprofiler/src/tools/tool.cpp @@ -627,7 +627,7 @@ ROCPROFILER_EXPORT bool OnLoad(void* table, uint64_t runtime_version, uint64_t f getFlushIntervalFromEnv(); getTracePeriodFromEnv(); - CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); + CHECK_ROCPROFILER(rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id)); bool want_pc_sampling = getenv("ROCPROFILER_PC_SAMPLING"); diff --git a/projects/rocprofiler/tests-v2/featuretests/profiler/profiler_gtest.cpp b/projects/rocprofiler/tests-v2/featuretests/profiler/profiler_gtest.cpp index 3cc9577b33..6797587c91 100644 --- a/projects/rocprofiler/tests-v2/featuretests/profiler/profiler_gtest.cpp +++ b/projects/rocprofiler/tests-v2/featuretests/profiler/profiler_gtest.cpp @@ -630,7 +630,7 @@ TEST_F(ATTCollection, WhenRunningATTItCollectsTraceData) { parameters.emplace_back(rocprofiler_att_parameter_t{ROCPROFILER_ATT_TOKEN_MASK2, 0xFFFF}); // create a session - result = rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id); + result = rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id); EXPECT_EQ(ROCPROFILER_STATUS_SUCCESS, result); // create a buffer to hold att trace records for each kernel launch @@ -742,7 +742,7 @@ TEST_F(ProfilerAPITest, WhenRunningMultipleThreadsProfilerAPIsWorkFine) { std::vector counters; counters.emplace_back("SQ_WAVES"); - CheckApi(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); + CheckApi(rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id)); rocprofiler_buffer_id_t buffer_id; CheckApi(rocprofiler_create_buffer(session_id, FlushCallback, 0x9999, &buffer_id)); @@ -907,7 +907,7 @@ TEST_F(ProfilerSPMTest, WhenRunningSPMItCollectsSPMData) { // spm_parameters.cpu_agent_id = NULL; spm_parameters.sampling_rate = 10000; // create a session - CheckApi(rocprofiler_create_session(ROCPROFILER_KERNEL_REPLAY_MODE, &session_id)); + CheckApi(rocprofiler_create_session(ROCPROFILER_NONE_REPLAY_MODE, &session_id)); // create a buffer to hold spm trace records for each kernel launch rocprofiler_buffer_id_t buffer_id; diff --git a/projects/rocprofiler/tests-v2/unittests/profiler_gtest.cpp b/projects/rocprofiler/tests-v2/unittests/profiler_gtest.cpp index 22ac271600..366e6abe23 100644 --- a/projects/rocprofiler/tests-v2/unittests/profiler_gtest.cpp +++ b/projects/rocprofiler/tests-v2/unittests/profiler_gtest.cpp @@ -320,29 +320,6 @@ TEST(WhenTestingTimeStampCollectionMode, TestSucceeds) { toolobj.DestroySession(session_id); } -TEST(WhenTestingApplicationReplayMode, TestSucceeds) { - std::vector counters; - counters.emplace_back("SQ_WAVES"); - rocprofiler_session_id_t session_id; - - rocprofiler::ROCProfiler_Singleton toolobj; - session_id = toolobj.CreateSession(ROCPROFILER_APPLICATION_REPLAY_MODE); - - rocprofiler_filter_id_t filter_id = - toolobj.GetSession(session_id) - ->CreateFilter(ROCPROFILER_COUNTERS_COLLECTION, - rocprofiler_filter_data_t{.counters_names = &counters[0]}, counters.size(), - rocprofiler_filter_property_t{}); - rocprofiler_buffer_id_t buffer_id = - toolobj.GetSession(session_id)->CreateBuffer(callback_fun, 0x8000); - toolobj.GetSession(session_id)->GetFilter(filter_id)->SetBufferId(buffer_id); - - rocprofiler::Session* session = toolobj.GetSession(session_id); - - EXPECT_TRUE(session->FindFilterWithKind(ROCPROFILER_COUNTERS_COLLECTION)); - toolobj.DestroySession(session_id); -} - TEST(WhenTrucatingLongKernelNames, KernelNameGetsTruncatedProperly) { std::string long_kernel_name = "void kernel_7r_3d_pml<32, 8, 4>(long long, long long, long long, int, "