ROCm/composable_kernel
1
0
Fork 0
mirror of https://github.com/ROCm/composable_kernel.git synced 2026-06-30 03:37:38 +00:00
Code Issues Packages Projects Releases Wiki Activity

clang-format

Browse Source 
Signed-off-by: TianyuanWu <Tianyuan.Wu@amd.com>
This commit is contained in:
TianyuanWu
2025-08-07 16:51:23 +08:00
parent dfb424d430
commit 91ee57f0df
11 changed files with 31 additions and 49 deletions
Download Patch File Download Diff File Expand all files Collapse all files

2
example/02_gemm_bilinear/gemm_bilinear_wmma_fp16.cpp
View File

@@ -21,7 +21,7 @@
struct AlphaBetaAdd
{
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta){};
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta) {};
template <typename E, typename C, typename D>
__host__ __device__ constexpr void operator()(E& e, const C& c, const D& d) const;

2
example/02_gemm_bilinear/gemm_bilinear_wmma_int8.cpp
View File

@@ -21,7 +21,7 @@
struct AlphaBetaAdd
{
AlphaBetaAdd(int alpha, int beta) : alpha_(alpha), beta_(beta){};
AlphaBetaAdd(int alpha, int beta) : alpha_(alpha), beta_(beta) {};
template <typename E, typename C, typename D>
__host__ __device__ constexpr void operator()(E& e, const C& c, const D& d) const;

2
example/02_gemm_bilinear/gemm_bilinear_xdl_fp16.cpp
View File

@@ -20,7 +20,7 @@
struct AlphaBetaAdd
{
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta){};
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta) {};
template <typename E, typename C, typename D>
__host__ __device__ constexpr void operator()(E& e, const C& c, const D& d) const;

2
example/60_gemm_multi_ABD/gemm_multi_ABD_xdl_fp16.cpp
View File

@@ -83,7 +83,7 @@ struct AddScale
struct AlphaBetaAdd
{
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta){};
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta) {};
template <typename E, typename C, typename D>
__host__ __device__ constexpr void operator()(E& e, const C& c, const D& d) const;

2
example/61_contraction_multi_ABD/contraction_multi_ABD_xdl_fp16.cpp
View File

@@ -42,7 +42,7 @@ static constexpr ck::index_t NumDimK = 2;
struct AlphaBetaAdd
{
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta){};
AlphaBetaAdd(float alpha, float beta) : alpha_(alpha), beta_(beta) {};
template <typename E, typename C, typename D>
__host__ __device__ constexpr void operator()(E& e, const C& c, const D& d) const;

8
include/ck/library/utility/host_tensor_generator.hpp
View File

@@ -483,7 +483,7 @@ struct GeneratorTensor_4
std::normal_distribution<float> distribution;
GeneratorTensor_4(float mean, float stddev, unsigned int seed = 1)
: generator(seed), distribution(mean, stddev){};
: generator(seed), distribution(mean, stddev) {};
template <typename... Is>
T operator()(Is...)
@@ -501,7 +501,7 @@ struct GeneratorTensor_4<ck::f4x2_pk_t>
std::normal_distribution<float> distribution;
GeneratorTensor_4(float mean, float stddev, unsigned int seed = 1)
: generator(seed), distribution(mean, stddev){};
: generator(seed), distribution(mean, stddev) {};
template <typename... Is>
ck::f4x2_pk_t operator()(Is...)
@@ -520,7 +520,7 @@ struct GeneratorTensor_4<ck::f6x32_pk_t>
std::normal_distribution<float> distribution;
GeneratorTensor_4(float mean, float stddev, unsigned int seed = 1)
: generator(seed), distribution(mean, stddev){};
: generator(seed), distribution(mean, stddev) {};
template <typename... Is>
ck::f6x32_pk_t operator()(Is...)
@@ -542,7 +542,7 @@ struct GeneratorTensor_4<ck::bf6x32_pk_t>
std::normal_distribution<float> distribution;
GeneratorTensor_4(float mean, float stddev, unsigned int seed = 1)
: generator(seed), distribution(mean, stddev){};
: generator(seed), distribution(mean, stddev) {};
template <typename... Is>
ck::bf6x32_pk_t operator()(Is...)

4
include/ck/tensor_operation/gpu/element/binary_element_wise_operation.hpp
View File

@@ -279,7 +279,7 @@ struct Subtract
struct Bilinear
{
Bilinear(float alpha = 1.f, float beta = 1.f) : alpha_(alpha), beta_(beta){};
Bilinear(float alpha = 1.f, float beta = 1.f) : alpha_(alpha), beta_(beta) {};
template <typename Y, typename X0, typename X1>
__host__ __device__ constexpr void operator()(Y&, const X0&, const X1&) const;
@@ -354,7 +354,7 @@ struct Bilinear
struct AddClamp
{
AddClamp(float floor = 0.f, float ceil = NumericLimits<float>::Max())
: floor_(floor), ceil_(ceil){};
: floor_(floor), ceil_(ceil) {};
template <typename Y, typename X0, typename X1>
__host__ __device__ constexpr void operator()(Y& y, const X0& x0, const X1& x1) const;

14
include/ck/tensor_operation/gpu/element/unary_element_wise_operation.hpp
View File

@@ -756,7 +756,7 @@ struct UnarySqrt
struct Clamp
{
Clamp(float floor = 0.f, float ceil = NumericLimits<float>::Max())
: floor_(floor), ceil_(ceil){};
: floor_(floor), ceil_(ceil) {};
template <typename Y, typename X>
__host__ __device__ constexpr void operator()(Y& y, const X& x) const;
@@ -1324,7 +1324,7 @@ struct Swish
struct SoftRelu
{
SoftRelu(float alpha = 1.f) : alpha_(alpha){};
SoftRelu(float alpha = 1.f) : alpha_(alpha) {};
template <typename T>
__host__ __device__ void operator()(T& y, const T& x) const
@@ -1353,7 +1353,7 @@ struct SoftRelu
struct Power
{
Power(float alpha = 0.f, float beta = 1.f, float gamma = 2.f)
: alpha_(alpha), beta_(beta), gamma_(gamma){};
: alpha_(alpha), beta_(beta), gamma_(gamma) {};
template <typename T>
__host__ __device__ void operator()(T& y, const T& x) const
@@ -1386,7 +1386,7 @@ struct Power
struct ClippedRelu
{
ClippedRelu(float alpha = 0.f, float beta = 1.f) : alpha_(alpha), beta_(beta){};
ClippedRelu(float alpha = 0.f, float beta = 1.f) : alpha_(alpha), beta_(beta) {};
template <typename T>
__host__ __device__ void operator()(T& y, const T& x) const
@@ -1415,7 +1415,7 @@ struct ClippedRelu
struct LeakyRelu
{
LeakyRelu(float alpha = 0.01f) : alpha_(alpha){};
LeakyRelu(float alpha = 0.01f) : alpha_(alpha) {};
template <typename T>
__host__ __device__ void operator()(T& y, const T& x) const
@@ -1442,7 +1442,7 @@ struct LeakyRelu
struct Elu
{
Elu(float alpha = 1.f) : alpha_(alpha){};
Elu(float alpha = 1.f) : alpha_(alpha) {};
template <typename T>
__host__ __device__ void operator()(T& y, const T& x) const
@@ -1469,7 +1469,7 @@ struct Elu
struct Logistic
{
Logistic(float alpha = 1.f) : alpha_(alpha){};
Logistic(float alpha = 1.f) : alpha_(alpha) {};
template <typename T>
__host__ __device__ void operator()(T& y, const T& x) const

2
include/ck_tile/core/config.hpp
View File

@@ -152,7 +152,7 @@
// buffer atomic add: floating point
#ifndef __HIP_DEVICE_COMPILE__ // for host code
#define CK_TILE_USE_AMD_BUFFER_ATOMIC_ADD_FLOAT 1
#elif defined(__gfx9__) || defined(__gfx12__)// for GPU code
#elif defined(__gfx9__) || defined(__gfx12__) // for GPU code
#define CK_TILE_USE_AMD_BUFFER_ATOMIC_ADD_FLOAT 1
#else // for GPU code
#define CK_TILE_USE_AMD_BUFFER_ATOMIC_ADD_FLOAT 0

12
include/ck_tile/ops/elementwise/unary_element_wise_operation.hpp
View File

@@ -1308,7 +1308,7 @@ struct Swish
struct SoftRelu
{
SoftRelu(float alpha = 1.f) : alpha_(alpha){};
SoftRelu(float alpha = 1.f) : alpha_(alpha) {};
template <typename T>
CK_TILE_HOST_DEVICE void operator()(T& y, const T& x) const
@@ -1327,7 +1327,7 @@ struct SoftRelu
struct Power
{
Power(float alpha = 0.f, float beta = 1.f, float gamma = 2.f)
: alpha_(alpha), beta_(beta), gamma_(gamma){};
: alpha_(alpha), beta_(beta), gamma_(gamma) {};
template <typename T>
CK_TILE_HOST_DEVICE void operator()(T& y, const T& x) const
@@ -1349,7 +1349,7 @@ struct Power
struct ClippedRelu
{
ClippedRelu(float alpha = 0.f, float beta = 1.f) : alpha_(alpha), beta_(beta){};
ClippedRelu(float alpha = 0.f, float beta = 1.f) : alpha_(alpha), beta_(beta) {};
template <typename T>
CK_TILE_HOST_DEVICE void operator()(T& y, const T& x) const
@@ -1368,7 +1368,7 @@ struct ClippedRelu
struct LeakyRelu
{
LeakyRelu(float alpha = 0.01f) : alpha_(alpha){};
LeakyRelu(float alpha = 0.01f) : alpha_(alpha) {};
template <typename T>
CK_TILE_HOST_DEVICE void operator()(T& y, const T& x) const
@@ -1385,7 +1385,7 @@ struct LeakyRelu
struct Elu
{
Elu(float alpha = 1.f) : alpha_(alpha){};
Elu(float alpha = 1.f) : alpha_(alpha) {};
template <typename T>
CK_TILE_HOST_DEVICE void operator()(T& y, const T& x) const
@@ -1402,7 +1402,7 @@ struct Elu
struct Logistic
{
Logistic(float alpha = 1.f) : alpha_(alpha){};
Logistic(float alpha = 1.f) : alpha_(alpha) {};
template <typename T>
CK_TILE_HOST_DEVICE void operator()(T& y, const T& x) const

30
include/ck_tile/ops/gemm/warp/warp_gemm_attribute_wmma_impl.hpp
View File

@@ -77,32 +77,14 @@ struct WarpGemmAttributeWmmaImpl
};
using DeviceIp = remove_cvref_t<decltype(ck_tile::get_device_arch())>;
using WarpGemmAttributeWmmaImpl_f32_16x16x16_f16_f16 = WarpGemmAttributeWmmaImpl<WmmaTraits<
DeviceIp,
fp16_t,
fp16_t,
float,
16,
16,
16>>;
using WarpGemmAttributeWmmaImpl_f32_16x16x16_f16_f16 =
WarpGemmAttributeWmmaImpl<WmmaTraits<DeviceIp, fp16_t, fp16_t, float, 16, 16, 16>>;
using WarpGemmAttributeWmmaImpl_f32_16x16x16_bf16_bf16 = WarpGemmAttributeWmmaImpl<WmmaTraits<
DeviceIp,
bf16_t,
bf16_t,
float,
16,
16,
16>>;
using WarpGemmAttributeWmmaImpl_f32_16x16x16_bf16_bf16 =
WarpGemmAttributeWmmaImpl<WmmaTraits<DeviceIp, bf16_t, bf16_t, float, 16, 16, 16>>;
using WarpGemmAttributeWmmaImpl_i32_16x16x16_i8_i8 = WarpGemmAttributeWmmaImpl<WmmaTraits<
DeviceIp,
int8_t,
int8_t,
int32_t,
16,
16,
16>>;
using WarpGemmAttributeWmmaImpl_i32_16x16x16_i8_i8 =
WarpGemmAttributeWmmaImpl<WmmaTraits<DeviceIp, int8_t, int8_t, int32_t, 16, 16, 16>>;
using WarpGemmAttributeWmmaImpl_f32_16x16x16_f8_f8 =
WarpGemmAttributeWmmaImpl<WmmaTraits<gfx12_t, fp8_t, fp8_t, float, 16, 16, 16>>;