Skip to content

Commit

Permalink
Adopt hipDataType and deprecate hipblasltDatatype_t
Browse files Browse the repository at this point in the history
  • Loading branch information
@jichangjichang
jichangjichang committed Jan 15, 2024
1 parent 4b3b344 commit 95131d6
Show file tree
Hide file tree
Showing 60 changed files with 2,127 additions and 1,939 deletions.
2 changes: 1 addition & 1 deletion CHANGELOG.md
View file
Original file line number Diff line number Diff line change
Expand Up @@ -27,7 +27,7 @@ Full documentation for hipBLASLt is available at [rocm.docs.amd.com/projects/hip

### Changes

* Replaced `hipblasDatatype_t` with `hipblasltDatatype_t`
* Replaced `hipblasDatatype_t` with `hipDataType`

### Removals

Expand Down
56 changes: 28 additions & 28 deletions clients/benchmarks/client.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -2,7 +2,7 @@
*
* MIT License
*
* Copyright (C) 2022-2023 Advanced Micro Devices, Inc.
* Copyright (C) 2022-2024 Advanced Micro Devices, Inc.
*
* Permission is hereby granted, free of charge, to any person obtaining a copy
* of this software and associated documentation files (the "Software"), to deal
Expand Down Expand Up @@ -68,8 +68,7 @@ void run_function(const func_map& map, const Arguments& arg, const std::string&
auto match = map.find(arg.function);
if(match == map.end())
throw std::invalid_argument("Invalid combination --function "s + arg.function
+ " --a_type "s + hipblaslt_datatype_to_string(arg.a_type)
+ msg);
+ " --a_type "s + hip_datatype_to_string(arg.a_type) + msg);
match->second(arg);
}

Expand All @@ -92,16 +91,17 @@ struct perf_matmul<
To,
Tc,
Tci,
std::enable_if_t<(std::is_same<TiA, hipblasLtHalf>{} && std::is_same<TiB, hipblasLtHalf>{})
|| (std::is_same<TiA, hip_bfloat16>{} && std::is_same<TiB, hip_bfloat16>{})
|| (std::is_same<TiA, float>{} && std::is_same<TiB, float>{})
|| (std::is_same<TiA, hipblaslt_f8>{} && std::is_same<TiB, hipblaslt_f8>{})
|| (std::is_same<TiA, hipblaslt_f8>{} && std::is_same<TiB, hipblaslt_bf8>{})
|| (std::is_same<TiA, hipblaslt_bf8>{} && std::is_same<TiB, hipblaslt_f8>{})
|| (std::is_same<TiA, double>{} && std::is_same<TiB, double>{})
|| (std::is_same<TiA, hipblasLtInt8>{} && std::is_same<TiB, hipblasLtInt8>{})
|| (std::is_same<TiA, hipblaslt_f8>{} && std::is_same<TiB, hipblasLtHalf>{})
|| (std::is_same<TiA, hipblasLtHalf>{} && std::is_same<TiB, hipblaslt_f8>{})>>
std::enable_if_t<
(std::is_same<TiA, hipblasLtHalf>{} && std::is_same<TiB, hipblasLtHalf>{})
|| (std::is_same<TiA, hip_bfloat16>{} && std::is_same<TiB, hip_bfloat16>{})
|| (std::is_same<TiA, float>{} && std::is_same<TiB, float>{})
|| (std::is_same<TiA, hipblaslt_f8_fnuz>{} && std::is_same<TiB, hipblaslt_f8_fnuz>{})
|| (std::is_same<TiA, hipblaslt_f8_fnuz>{} && std::is_same<TiB, hipblaslt_bf8_fnuz>{})
|| (std::is_same<TiA, hipblaslt_bf8_fnuz>{} && std::is_same<TiB, hipblaslt_f8_fnuz>{})
|| (std::is_same<TiA, double>{} && std::is_same<TiB, double>{})
|| (std::is_same<TiA, hipblasLtInt8>{} && std::is_same<TiB, hipblasLtInt8>{})
|| (std::is_same<TiA, hipblaslt_f8_fnuz>{} && std::is_same<TiB, hipblasLtHalf>{})
|| (std::is_same<TiA, hipblasLtHalf>{} && std::is_same<TiB, hipblaslt_f8_fnuz>{})>>
: hipblaslt_test_valid
{
void operator()(const Arguments& arg)
Expand Down Expand Up @@ -731,38 +731,38 @@ try
}

std::transform(precision.begin(), precision.end(), precision.begin(), ::tolower);
auto prec = string_to_hipblaslt_datatype(precision);
if(prec == static_cast<hipblasltDatatype_t>(0))
auto prec = string_to_hip_datatype(precision);
if(prec == HIPBLASLT_DATATYPE_INVALID)
throw std::invalid_argument("Invalid value for --precision " + precision);

arg.a_type = a_type == "" ? prec : string_to_hipblaslt_datatype(a_type);
if(arg.a_type == static_cast<hipblasltDatatype_t>(0))
arg.a_type = a_type == "" ? prec : string_to_hip_datatype(a_type);
if(arg.a_type == HIPBLASLT_DATATYPE_INVALID)
throw std::invalid_argument("Invalid value for --a_type " + a_type);

arg.b_type = b_type == "" ? prec : string_to_hipblaslt_datatype(b_type);
if(arg.b_type == static_cast<hipblasltDatatype_t>(0))
arg.b_type = b_type == "" ? prec : string_to_hip_datatype(b_type);
if(arg.b_type == HIPBLASLT_DATATYPE_INVALID)
throw std::invalid_argument("Invalid value for --b_type " + b_type);

arg.c_type = c_type == "" ? prec : string_to_hipblaslt_datatype(c_type);
if(arg.c_type == static_cast<hipblasltDatatype_t>(0))
arg.c_type = c_type == "" ? prec : string_to_hip_datatype(c_type);
if(arg.c_type == HIPBLASLT_DATATYPE_INVALID)
throw std::invalid_argument("Invalid value for --c_type " + c_type);

arg.d_type = d_type == "" ? prec : string_to_hipblaslt_datatype(d_type);
if(arg.d_type == static_cast<hipblasltDatatype_t>(0))
arg.d_type = d_type == "" ? prec : string_to_hip_datatype(d_type);
if(arg.d_type == HIPBLASLT_DATATYPE_INVALID)
throw std::invalid_argument("Invalid value for --d_type " + d_type);

bool is_f16 = arg.a_type == HIPBLASLT_R_16F || arg.a_type == HIPBLASLT_R_16B;
bool is_f32 = arg.a_type == HIPBLASLT_R_32F;
bool is_f16 = arg.a_type == HIP_R_16F || arg.a_type == HIP_R_16BF;
bool is_f32 = arg.a_type == HIP_R_32F;
arg.compute_type = compute_type == "" ? (HIPBLASLT_COMPUTE_F32)
: string_to_hipblaslt_computetype(compute_type);
if(arg.compute_type == static_cast<hipblasLtComputeType_t>(0))
throw std::invalid_argument("Invalid value for --compute_type " + compute_type);

if(string_to_hipblaslt_datatype(bias_type) == static_cast<hipblasltDatatype_t>(0)
&& bias_type != "" && bias_type != "default")
if(string_to_hip_datatype(bias_type) == HIPBLASLT_DATATYPE_INVALID && bias_type != ""
&& bias_type != "default")
throw std::invalid_argument("Invalid value for --bias_type " + bias_type);
else
arg.bias_type = string_to_hipblaslt_datatype(bias_type);
arg.bias_type = string_to_hip_datatype(bias_type);

arg.initialization = string2hipblaslt_initialization(initialization);
if(arg.initialization == static_cast<hipblaslt_initialization>(0))
Expand Down
137 changes: 76 additions & 61 deletions clients/benchmarks/client_extop_amax.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -2,7 +2,7 @@
*
* MIT License
*
* Copyright (C) 2023 Advanced Micro Devices, Inc.
* Copyright (C) 2023-2024 Advanced Micro Devices, Inc.
*
* Permission is hereby granted, free of charge, to any person obtaining a copy
* of this software and associated documentation files (the "Software"), to deal
Expand All @@ -24,81 +24,92 @@
*
*******************************************************************************/

#include <numeric>
#include <iostream>
#include <vector>
#include <random>
#include <type_traits>
#include <hip/hip_runtime.h>
#include <hip/hip_runtime_api.h>
#include <hipblaslt/hipblaslt.h>
#include <hipblaslt/hipblaslt-ext-op.h>
#include <hipblaslt/hipblaslt.h>
#include <hipblaslt_datatype2string.hpp>
#include <hipblaslt_init.hpp>
#include <iostream>
#include <numeric>
#include <random>
#include <type_traits>
#include <vector>

void printUsage(char *programName) {
std::cout
<< "Usage: " << programName << " <options>\n"
<< "options:\n"
<< "\t-h, --help\t\t\tShow this help message\n"
<< "\t-t, --type\t\t\tType of problem, default is S.\n"
<< "\t-d, --dtype\t\t\tDest Type of problem, default is S.\n"
<< "\t-m, --m\t\t\t\tSize of dim 0, default is 64\n"
<< "\t-n, --n\t\t\t\tSize of dim 1, default is 64\n"
<< "\t--initialization \t\tInitialize matrix data. Options: rand_int, trig_float, "
void printUsage(char* programName)
{
std::cout << "Usage: " << programName << " <options>\n"
<< "options:\n"
<< "\t-h, --help\t\t\tShow this help message\n"
<< "\t-t, --type\t\t\tType of problem, default is S.\n"
<< "\t-d, --dtype\t\t\tDest Type of problem, default is S.\n"
<< "\t-m, --m\t\t\t\tSize of dim 0, default is 64\n"
<< "\t-n, --n\t\t\t\tSize of dim 1, default is 64\n"
<< "\t--initialization \t\tInitialize matrix data. Options: rand_int, trig_float, "
"hpl(floating). (default is hpl)\n";
}

template<typename T>
template <typename T>
T abs(T a)
{
return (a > 0) ? a : -a;
return (a > 0) ? a : -a;
}

template<typename T>
template <typename T>
T max(T a, T b)
{
return (a > b) ? a : b;
}

template<typename Ti, typename To>
void cpuAMax(To *out, Ti *in, std::uint32_t length)
template <typename Ti, typename To>
void cpuAMax(To* out, Ti* in, std::uint32_t length)
{
// calculate amax
Ti m = 0;
for(int j=0; j<length; j++) {
for(int j = 0; j < length; j++)
{
m = max(m, abs(in[j]));
}
out[0] = To(m);
}

int parseArgs(int argc, char **argv, std::string& type, std::string& dtype, size_t &m, size_t &n, hipblaslt_initialization& init)
int parseArgs(int argc,
char** argv,
std::string& type,
std::string& dtype,
size_t& m,
size_t& n,
hipblaslt_initialization& init)
{
if (argc <= 1)
if(argc <= 1)
{
return EXIT_SUCCESS;
}

for (int i = 1; i < argc; ++i)
for(int i = 1; i < argc; ++i)
{
std::string arg = argv[i];

if ((arg.at(0) == '-') || ((arg.at(0) == '-') && (arg.at(1) == '-')))
if((arg.at(0) == '-') || ((arg.at(0) == '-') && (arg.at(1) == '-')))
{
if((arg == "-h") || (arg == "--help"))
{
return EXIT_FAILURE;
}
else if (arg == "-t" || arg == "--type") {
else if(arg == "-t" || arg == "--type")
{
type = argv[++i];
}
else if (arg == "-d" || arg == "--dtype") {
else if(arg == "-d" || arg == "--dtype")
{
dtype = argv[++i];
}
else if (arg == "-m" || arg == "--m") {
else if(arg == "-m" || arg == "--m")
{
n = std::stoul(argv[++i]);
}
else if (arg == "-n" || arg == "--n") {
else if(arg == "-n" || arg == "--n")
{
n = std::stoul(argv[++i]);
}
else if(arg == "--initialization" || arg == "--init")
Expand Down Expand Up @@ -129,7 +140,8 @@ template <typename Dtype>
void dumpBuffer(const char* title, Dtype* data, int N)
{
std::cout << "----- " << title << "----- " << std::endl;
for(int n=0; n<N; n++) {
for(int n = 0; n < N; n++)
{
std::cout << float(data[n]) << " ";
}
std::cout << std::endl;
Expand All @@ -140,8 +152,9 @@ template <typename T>
void compare(const char* title, const std::vector<T>& cpuOutput, const std::vector<T>& refOutput)
{
T maxErr = 0.0;
for (int i=0; i<cpuOutput.size(); i++) {
T err = abs(refOutput[i] - cpuOutput[i]);
for(int i = 0; i < cpuOutput.size(); i++)
{
T err = abs(refOutput[i] - cpuOutput[i]);
maxErr = max(maxErr, err);
}

Expand Down Expand Up @@ -170,18 +183,18 @@ void initData(DType* data, std::size_t numElements, hipblaslt_initialization ini
}
}

template<typename Ti, typename To>
int AmaxTest(hipblasltDatatype_t type, hipblasltDatatype_t dtype, int m, int n, hipblaslt_initialization& init)
template <typename Ti, typename To>
int AmaxTest(hipDataType type, hipDataType dtype, int m, int n, hipblaslt_initialization& init)
{
int numElements = m * n;
std::size_t tiNumBytes = sizeof(Ti);
std::size_t toNumBytes = sizeof(To);
int numElements = m * n;
std::size_t tiNumBytes = sizeof(Ti);
std::size_t toNumBytes = sizeof(To);

To *gpuOutput{nullptr};
Ti *gpuInput{nullptr};
To* gpuOutput{nullptr};
Ti* gpuInput{nullptr};

auto hipErr = hipMalloc(&gpuOutput, toNumBytes);
hipErr = hipMalloc(&gpuInput, m * n * tiNumBytes);
hipErr = hipMalloc(&gpuInput, m * n * tiNumBytes);

std::vector<To> cpuOutput(1, 0.f);
std::vector<Ti> cpuInput(m * n, 0.f);
Expand All @@ -207,12 +220,13 @@ int AmaxTest(hipblasltDatatype_t type, hipblasltDatatype_t dtype, int m, int n,
compare("Output", cpuOutput, refOutput);

hipEvent_t beg, end;
hipErr = hipEventCreate(&beg);
hipErr = hipEventCreate(&end);
hipErr = hipEventCreate(&beg);
hipErr = hipEventCreate(&end);
int numRuns = 200;
hipErr = hipEventRecord(beg, stream);
hipErr = hipEventRecord(beg, stream);

for (int i = 0; i < numRuns; ++i) {
for(int i = 0; i < numRuns; ++i)
{
hipblasltErr = hipblasltExtAMax(type, dtype, gpuOutput, gpuInput, m, n, stream);
}
hipErr = hipEventRecord(end, stream);
Expand All @@ -232,27 +246,28 @@ int AmaxTest(hipblasltDatatype_t type, hipblasltDatatype_t dtype, int m, int n,
return 0;
}


int main(int argc, char **argv) {
std::string type{"S"};
std::string dtype{"S"};
std::size_t m{64};
std::size_t n{64};
int main(int argc, char** argv)
{
std::string type{"S"};
std::string dtype{"S"};
std::size_t m{64};
std::size_t n{64};
hipblaslt_initialization init{hipblaslt_initialization::hpl};

if (auto err = parseArgs(argc, argv, type, dtype, m, n, init)) {
if(auto err = parseArgs(argc, argv, type, dtype, m, n, init))
{
printUsage(argv[0]);
return err;
}

if ((type == "S" || type == "s") && (type == dtype))
return AmaxTest<float, float>(HIPBLASLT_R_32F, HIPBLASLT_R_32F, m, n, init);
else if ((type == "S" || type == "s") && (dtype == "H" || dtype == "H"))
return AmaxTest<float, hipblasLtHalf>(HIPBLASLT_R_32F, HIPBLASLT_R_16F, m, n, init);
else if ((type == "H" || type == "h") && (type == dtype))
return AmaxTest<hipblasLtHalf, hipblasLtHalf>(HIPBLASLT_R_16F, HIPBLASLT_R_16F, m, n, init);
else if ((type == "H" || type == "h") && (dtype == "S" || dtype == "s"))
return AmaxTest<hipblasLtHalf, float>(HIPBLASLT_R_16F, HIPBLASLT_R_32F, m, n, init);
if((type == "S" || type == "s") && (type == dtype))
return AmaxTest<float, float>(HIP_R_32F, HIP_R_32F, m, n, init);
else if((type == "S" || type == "s") && (dtype == "H" || dtype == "H"))
return AmaxTest<float, hipblasLtHalf>(HIP_R_32F, HIP_R_16F, m, n, init);
else if((type == "H" || type == "h") && (type == dtype))
return AmaxTest<hipblasLtHalf, hipblasLtHalf>(HIP_R_16F, HIP_R_16F, m, n, init);
else if((type == "H" || type == "h") && (dtype == "S" || dtype == "s"))
return AmaxTest<hipblasLtHalf, float>(HIP_R_16F, HIP_R_32F, m, n, init);
else
std::cout << "Unsupported data type " << type << std::endl;

Expand Down
17 changes: 4 additions & 13 deletions clients/benchmarks/client_extop_layernorm.cpp
View file
Original file line number Diff line number Diff line change
Expand Up @@ -2,7 +2,7 @@
*
* MIT License
*
* Copyright (C) 2023 Advanced Micro Devices, Inc.
* Copyright (C) 2023-2024 Advanced Micro Devices, Inc.
*
* Permission is hereby granted, free of charge, to any person obtaining a copy
* of this software and associated documentation files (the "Software"), to deal
Expand Down Expand Up @@ -284,17 +284,8 @@ int main(int argc, char** argv)
hipStream_t stream{};
hipErr = hipStreamCreate(&stream);
//warmup
auto hipblasltErr = hipblasltExtLayerNorm(HIPBLASLT_R_32F,
gpuOutput,
gpuMean,
gpuInvvar,
gpuInput,
m,
n,
1e-05,
gpuGamma,
gpuBeta,
stream);
auto hipblasltErr = hipblasltExtLayerNorm(
HIP_R_32F, gpuOutput, gpuMean, gpuInvvar, gpuInput, m, n, 1e-05, gpuGamma, gpuBeta, stream);

hipErr = hipMemcpyDtoH(cpuOutput.data(), gpuOutput, numElements * elementNumBytes);
hipErr = hipMemcpyDtoH(cpuMean.data(), gpuMean, m * elementNumBytes);
Expand Down Expand Up @@ -325,7 +316,7 @@ int main(int argc, char** argv)

for(int i = 0; i < numRuns; ++i)
{
hipblasltErr = hipblasltExtLayerNorm(HIPBLASLT_R_32F,
hipblasltErr = hipblasltExtLayerNorm(HIP_R_32F,
gpuOutput,
gpuMean,
gpuInvvar,
Expand Down
Loading

0 comments on commit 95131d6

Please sign in to comment.