AI/rocm-systems
2
0
派生 0
代码 工单 合并请求 工作流 软件包 项目 发布 百科 活动

SWDEV-405920: V2 clean up for unimplemented and corrupted APIs

浏览代码 
Change-Id: I2c88d8415a3a386377e99b93295adb82e1f7195e


[ROCm/rocprofiler commit: 37c48ce0ab]
这个提交包含在：
Ammar ELWazir
2023-06-20 20:51:47 +00:00
提交者 Ammar Elwazir
父节点 6964105c1f
当前提交 eb769499b6
共修改 15 个文件，包含 11 行新增和 408 行删除
下载 Patch 文件 下载 Diff 文件 展开所有文件 折叠所有文件
projects/rocprofiler/CHANGELOG.md
+1
查看文件
@@ -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 <as.s> --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
projects/rocprofiler/include/rocprofiler/v2/rocprofiler.h
-84
查看文件
@@ -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
projects/rocprofiler/samples/CMakeLists.txt
-27
查看文件
@@ -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})
projects/rocprofiler/samples/counters_sampler/pcie_counters_example.cpp
+3 -3
查看文件
@@ -3,7 +3,7 @@
int main(int argc, char** argv) {
int* gpuMem;
int counter_option = 0;
std::vector<std::string> 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<char*>(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,
projects/rocprofiler/samples/counters_sampler/xgmi_counters_sampler_example.cpp
+1 -1
查看文件
@@ -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;
projects/rocprofiler/samples/profiler/application_replay_sample.cpp
-72
查看文件
@@ -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<const char*> 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;
}
projects/rocprofiler/samples/profiler/kernel_replay_sample.cpp
-68
查看文件
@@ -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<const char*> 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;
}
projects/rocprofiler/samples/profiler/user_replay_sample.cpp
-78
查看文件
@@ -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<const char*> 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;
}
projects/rocprofiler/samples/tracer/sample.cpp
+1 -1
查看文件
@@ -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;
projects/rocprofiler/samples/tracer/sample_async.cpp
+1 -1
查看文件
@@ -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;
projects/rocprofiler/src/api/exportmap
-4
查看文件
@@ -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;
projects/rocprofiler/src/api/rocprofilerv2.cpp
-42
查看文件
@@ -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) {
projects/rocprofiler/src/tools/tool.cpp
+1 -1
查看文件
@@ -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");
projects/rocprofiler/tests-v2/featuretests/profiler/profiler_gtest.cpp
+3 -3
查看文件
@@ -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<const char*> 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;
projects/rocprofiler/tests-v2/unittests/profiler_gtest.cpp
-23
查看文件
@@ -320,29 +320,6 @@ TEST(WhenTestingTimeStampCollectionMode, TestSucceeds) {
toolobj.DestroySession(session_id);
}
TEST(WhenTestingApplicationReplayMode, TestSucceeds) {
std::vector<const char*> 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, "