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

Swizzle support for hipBLASLt #1584

Merged
merged 23 commits into from
Feb 6, 2025
Merged

Swizzle support for hipBLASLt #1584

Show file tree
Hide file tree
Changes from 20 commits
Commits
Show all changes
23 commits
Select commit Hold shift + click to select a range
adc00c8
Initial draft for hipBLASLt tensor swizzling
Serge45 Nov 21, 2024
c6c87bd
Refined tensor-swizzling example
Serge45 Nov 22, 2024
61cd9c5
Fixed incorrect check of opB and swizzleB
Serge45 Nov 22, 2024
de80157
Draft version of tensor swizzling A for hipblaslt-bench
Serge45 Nov 27, 2024
104a77a
Added free size logic yamls for swizzle-a for demostration
Serge45 Nov 27, 2024
16df6ad
hipblaslt-bench: copy init buffer back to CPU for swizzle case
jichangjichang Dec 6, 2024
97f7c42
update siwzzleA HHS/HSS
jichangjichang Dec 6, 2024
4357cbf
clean BiasDataTypeList from swizzle pure gemm yaml
jichangjichang Dec 6, 2024
d6c0cd6
update AF0EM,AF1EM and ASEM into swizzle logic yaml
jichangjichang Dec 26, 2024
ad05e63
Added C++ getHeuristic sample for swizzle-A
Serge45 Dec 26, 2024
98521a1
update HSS STA freesize
jichangjichang Jan 6, 2025
eaa3578
Amended swizzle example with padding
Serge45 Jan 4, 2025
a0807e8
Added weight swizzle and pad example
Serge45 Jan 4, 2025
23c0aec
Fixed auto-padding in hipblaslt-bench
Serge45 Jan 4, 2025
3c0ac9c
Updated shared headers in clients folder
Serge45 Jan 20, 2025
ee47f83
Renamed enum for swizzled tensor
Serge45 Jan 20, 2025
3915c3d
Make swizzleB in hipBLASLt API properly propagate
Serge45 Jan 21, 2025
27f23b3
Amended document of swizzle for hipBLASLT API
Serge45 Jan 22, 2025
43e3c09
Fixed typo in hipblaslt.h
Serge45 Jan 23, 2025
f581699
Added swizzle enums and updated API reference
Serge45 Jan 23, 2025
0675f82
Added hipblasLtOrder_t to datatypes doc
Serge45 Jan 23, 2025
292a4ec
Updated logics for swizzle-A kernels
Serge45 Jan 23, 2025
8108879
Update docs/api-reference.rst
Serge45 Jan 24, 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
10 changes: 10 additions & 0 deletions clients/benchmarks/client.cpp
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -413,6 +413,10 @@ try
value<char>(&arg.transB)->default_value('N'),
"N = no transpose, T = transpose")

("swizzleA",
value<bool>(&arg.swizzle_a)->default_value(false),
"Enable tensor swizzling for A")

("batch_count",
value<int32_t>(&arg.batch_count)->default_value(1),
"Number of matrices. Only applicable to batched and strided_batched routines")
Expand Down Expand Up @@ -670,6 +674,12 @@ try
return 1;
}

if(arg.swizzle_a && (arg.transA != 'T' || arg.transB != 'N' || a_type != "f16_r"))
{
hipblaslt_cerr << "For swizzle-A, problem type must be FP16 TN" << std::endl;
return 1;
}

// transfer local variable state
ArgumentModel_set_log_function_name(log_function_name);

Expand Down
Loading
Loading