mirror of
https://github.com/ROCm/composable_kernel.git
synced 2026-05-14 18:17:44 +00:00
* fix relu
* clean up
* clean up
* adding 1x1 conv
* adding 1x1 conv
* added 1x1 conv
* refactor
* refactor
* refactor
* added profiler for conv+bias+relu+add
* clean up
* adding conv+bias+relu
* adding conv+bias+relu
* added conv+bias+relu
* Update README.md
* update cpu verification
* adding c shuffle
* update static_tensor for dealing with invalid element
* adding c shuffle
* debugging
* fix bug
* convert to fp16 before shuffle
* shuffle more than one M/NRepeat
* clean up
* remove coordinate step hack from GridwiseGemm_k0mk1_k0nk1_mn_xdlops_v3r1
* clean up
* remove coordinate step hack from all gridwise gemm xdl
* clean up coordinate step hack
* clean up coordinate step hack
* ThreadwiseTensorSliceTransfer_v3r2 support pointwise op on both src and dst
* adding output shuffle in conv+bias+relu+add
* update
* added conv+bias+relu+add with c shuffle
* added conv+bias+relu+add with c shuffle
* fix forward_sweep bugs in threadwise copy
* clean up
* refactor
* clean up
* clean up
* added conv_c_shuffle+bias_relu
* clean up
* added conv+bias+relu+atomic_add
* clean up
* clean up
* clean up
* clean up
* clean up
* clean up
* misc fixes; add 1x1 specialization
* clean up
* delete unused device op
* clean up
* add support for odd C value
[ROCm/composable_kernel commit: acbd7bd7c5]
47 lines
1.4 KiB
C++
47 lines
1.4 KiB
C++
#include <iostream>
|
|
#include <numeric>
|
|
#include <initializer_list>
|
|
#include <cstdlib>
|
|
#include <stdlib.h>
|
|
#include <half.hpp>
|
|
|
|
int profile_gemm(int, char*[]);
|
|
int profile_conv_fwd(int, char*[]);
|
|
int profile_conv_fwd_bias_relu(int, char*[]);
|
|
int profile_conv_fwd_bias_relu_add(int, char*[]);
|
|
int profile_conv_fwd_bias_relu_atomic_add(int, char*[]);
|
|
|
|
int main(int argc, char* argv[])
|
|
{
|
|
if(strcmp(argv[1], "gemm") == 0)
|
|
{
|
|
return profile_gemm(argc, argv);
|
|
}
|
|
else if(strcmp(argv[1], "conv_fwd") == 0)
|
|
{
|
|
return profile_conv_fwd(argc, argv);
|
|
}
|
|
else if(strcmp(argv[1], "conv_fwd_bias_relu") == 0)
|
|
{
|
|
return profile_conv_fwd_bias_relu(argc, argv);
|
|
}
|
|
else if(strcmp(argv[1], "conv_fwd_bias_relu_add") == 0)
|
|
{
|
|
return profile_conv_fwd_bias_relu_add(argc, argv);
|
|
}
|
|
else if(strcmp(argv[1], "conv_fwd_bias_relu_atomic_add") == 0)
|
|
{
|
|
return profile_conv_fwd_bias_relu_atomic_add(argc, argv);
|
|
}
|
|
else
|
|
{
|
|
printf("arg1: tensor operation (gemm: GEMM;\n"
|
|
" conv_fwd: ForwardConvolution;\n"
|
|
" conv_fwd_bias_relu: ForwardConvolution+Bias+ReLU)\n"
|
|
" conv_fwd_bias_relu_add: ForwardConvolution+Bias+ReLU+Add)\n"
|
|
" conv_fwd_bias_relu_atomic_add: "
|
|
"ForwardConvolution+Bias+ReLU+AtomicAdd)\n");
|
|
return 0;
|
|
}
|
|
}
|