Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Add tensilelite client performance args to hipblaslt-bench #1544

Draft
wants to merge 2 commits into
base: develop
Choose a base branch
from
+867 −77
Draft

Add tensilelite client performance args to hipblaslt-bench #1544

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
f32d962
Add tensilelite clients to hipblaslt-bench
NaveenElumalaiAMD Jan 13, 2025
b26eabc
More changes
NaveenElumalaiAMD Jan 13, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
2 changes: 1 addition & 1 deletion clients/CMakeLists.txt
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -105,7 +105,7 @@ if( BUILD_CLIENTS_BENCHMARKS OR BUILD_CLIENTS_TESTS)
set( hipblaslt_test_bench_common
../common/singletons.cpp
../common/utility.cpp
../common/frequency_monitor.cpp
../common/performance_monitor.cpp
../common/cblas_interface.cpp
../common/argument_model.cpp
../common/hipblaslt_parse_data.cpp
Expand Down
10 changes: 5 additions & 5 deletions clients/benchmarks/client.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -39,7 +39,7 @@
#include <string>
#include <type_traits>

#include "frequency_monitor.hpp"
#include "performance_monitor.hpp"

#include "testing_matmul.hpp"

Expand All @@ -50,7 +50,7 @@
using namespace roc; // For emulated program_options
using namespace std::literals; // For std::string literals of form "str"s

struct perf_matmul: hipblaslt_test_valid
struct perf_matmul : hipblaslt_test_valid
{
void operator()(const Arguments& arg)
{
Expand Down Expand Up @@ -741,8 +741,8 @@ try
throw std::invalid_argument("Invalid Device ID");
set_device(device_id);

FrequencyMonitor& freq_monitor = getFrequencyMonitor();
freq_monitor.set_device_id(device_id);
PerformanceMonitor& perf_monitor = getPerformanceMonitor();
perf_monitor.set_device_id(device_id);

if(datafile)
return hipblaslt_bench_datafile(filter, any_stride);
Expand Down Expand Up @@ -873,7 +873,7 @@ try

arg.norm_check_assert = false;
int status = run_bench_test(arg, filter, any_stride);
freeFrequencyMonitor();
freePerformanceMonitor();
return status;
}
catch(const std::invalid_argument& exp)
Expand Down
50 changes: 39 additions & 11 deletions clients/common/argument_model.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -25,7 +25,7 @@
*******************************************************************************/

#include "argument_model.hpp"
#include "frequency_monitor.hpp"
#include "performance_monitor.hpp"

// this should have been a member variable but due to the complex variadic template this singleton allows global control

Expand All @@ -41,31 +41,59 @@ bool ArgumentModel_get_log_function_name()
return log_function_name;
}

void ArgumentModel_log_frequencies(hipblaslt_internal_ostream& name_line,
void ArgumentModel_log_performance(hipblaslt_internal_ostream& name_line,
hipblaslt_internal_ostream& val_line)
{

FrequencyMonitor& frequency_monitor = getFrequencyMonitor();
if(!frequency_monitor.enabled())
PerformanceMonitor& performance_monitor = getPerformanceMonitor();
if(!performance_monitor.enabled())
return;
if(!frequency_monitor.detailedReport())

name_line << ",Total Granularity";
val_line << "," << performance_monitor.getTotalGranularityValue();

name_line << ",Tiles Per-CU";
val_line << "," << performance_monitor.getTilesPerCuValue();

name_line << ",Tile-0 Granularity";
val_line << "," << performance_monitor.getTile0Granularity();

name_line << ",Tile-1 Granularity";
val_line << "," << performance_monitor.getTile1Granularity();

name_line << ",CU granularity";
val_line << "," << performance_monitor.getCuGranularity();

name_line << ",Wave granularity";
val_line << "," << performance_monitor.getWaveGranularity();

name_line << ",#CU's";
val_line << "," << performance_monitor.getCUs();

name_line << ",mem-read-bytes";
val_line << "," << performance_monitor.getMemReadBytes();

name_line << ",mem-write-bytes";
val_line << "," << performance_monitor.getMemWriteBytesD();

if(!performance_monitor.detailedReport())
{
name_line << ",lowest-avg-freq";
val_line << "," << frequency_monitor.getLowestAverageSYSCLK();
val_line << "," << performance_monitor.getLowestAverageSYSCLK();

name_line << ",lowest-median-freq";
val_line << "," << frequency_monitor.getLowestMedianSYSCLK();
val_line << "," << performance_monitor.getLowestMedianSYSCLK();
}
else
{
auto allAvgSYSCLK = frequency_monitor.getAllAverageSYSCLK();
auto allAvgSYSCLK = performance_monitor.getAllAverageSYSCLK();
for(int i = 0; i < allAvgSYSCLK.size(); i++)
{
name_line << ",avg-freq_" << i;
val_line << "," << allAvgSYSCLK[i];
}

auto allMedianSYSCLK = frequency_monitor.getAllMedianSYSCLK();
auto allMedianSYSCLK = performance_monitor.getAllMedianSYSCLK();
for(int i = 0; i < allMedianSYSCLK.size(); i++)
{
name_line << ",median-freq_" << i;
Expand All @@ -74,8 +102,8 @@ void ArgumentModel_log_frequencies(hipblaslt_internal_ostream& name_line,
}

name_line << ",avg-MCLK";
val_line << "," << frequency_monitor.getAverageMEMCLK();
val_line << "," << performance_monitor.getAverageMEMCLK();

name_line << ",median-MCLK";
val_line << "," << frequency_monitor.getMedianMEMCLK();
val_line << "," << performance_monitor.getMedianMEMCLK();
}
Loading