GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ > Struct Template Reference

GroupedConvBwdWeightKernelArgs&lt; GroupedConvTraitsType_ &gt; Struct Template Reference#

Composable Kernel: ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ > Struct Template Reference
Public Types | Public Member Functions | Public Attributes | Static Public Attributes | List of all members
ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ > Struct Template Reference

The Grouped Convolution kernel device arguments. More...

#include <grouped_convolution_backward_weight_kernel.hpp>

Public Types

using ConvToGemmTransformer
using ABCGridDescs
using AGridDescKM = remove_cvref_t<decltype(ABCGridDescs{}[number<0>{}])>
using BGridDescKN = remove_cvref_t<decltype(ABCGridDescs{}[number<1>{}])>
using CGridDescMN = remove_cvref_t<decltype(ABCGridDescs{}[number<2>{}])>

Public Member Functions

template<typename InLay = typename GroupedConvTraitsType_::InLayout, typename WeiLay = typename GroupedConvTraitsType_::WeiLayout, typename OutLay = typename GroupedConvTraitsType_::OutLayout, typename std::enable_if< std::is_same_v< InLay, tensor_layout::convolution::NWGC > &&std::is_same_v< WeiLay, tensor_layout::convolution::GKXC > &&std::is_same_v< OutLay, tensor_layout::convolution::NWGK >, bool >::type = false>
CK_TILE_HOST GroupedConvBwdWeightKernelArgs (const GroupedConvBwdWeightHostArgs &args)
template<typename InLay = typename GroupedConvTraitsType_::InLayout, typename WeiLay = typename GroupedConvTraitsType_::WeiLayout, typename OutLay = typename GroupedConvTraitsType_::OutLayout, typename std::enable_if< std::is_same_v< InLay, tensor_layout::convolution::NHWGC > &&std::is_same_v< WeiLay, tensor_layout::convolution::GKYXC > &&std::is_same_v< OutLay, tensor_layout::convolution::NHWGK >, bool >::type = false>
CK_TILE_HOST GroupedConvBwdWeightKernelArgs (const GroupedConvBwdWeightHostArgs &args)
template<typename InLay = typename GroupedConvTraitsType_::InLayout, typename WeiLay = typename GroupedConvTraitsType_::WeiLayout, typename OutLay = typename GroupedConvTraitsType_::OutLayout, typename std::enable_if< std::is_same_v< InLay, tensor_layout::convolution::NDHWGC > &&std::is_same_v< WeiLay, tensor_layout::convolution::GKZYXC > &&std::is_same_v< OutLay, tensor_layout::convolution::NDHWGK >, bool >::type = false>
CK_TILE_HOST GroupedConvBwdWeightKernelArgs (const GroupedConvBwdWeightHostArgs &args)

Public Attributes

array< index_t, NonSpatialDims+GroupedConvTraitsType_::NDimSpatial > in_g_n_c_wis_lengths
array< index_t, NonSpatialDims+GroupedConvTraitsType_::NDimSpatial > wei_g_k_c_xs_lengths
array< index_t, NonSpatialDims+GroupedConvTraitsType_::NDimSpatial > out_g_n_k_wos_lengths
array< index_t, GroupedConvTraitsType_::NDimSpatial > conv_filter_strides
array< index_t, GroupedConvTraitsType_::NDimSpatial > conv_filter_dilations
array< index_t, GroupedConvTraitsType_::NDimSpatial > input_left_pads
array< index_t, GroupedConvTraitsType_::NDimSpatial > input_right_pads
index_t k_batch
index_t GemmM
index_t GemmN
index_t GemmK
index_t GemmBatch
index_t NumGroupsPerBatch
const void * out_ptr
const void * in_ptr
std::array< const void *, NumDTensor > ds_ptr
void * wei_ptr
AGridDescKM a_grid_desc_k_m
BGridDescKN b_grid_desc_k_n
CGridDescMN c_grid_desc_m_n
long_index_t group_stride_a
long_index_t group_stride_b
long_index_t group_stride_c

Static Public Attributes

static constexpr index_t NumDTensor = GroupedConvTraitsType_::NumDTensor
static constexpr index_t NonSpatialDims = 3

Detailed Description

template<typename GroupedConvTraitsType_>
struct ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >

The Grouped Convolution kernel device arguments.

Member Typedef Documentation

◆ ABCGridDescs

template<typename GroupedConvTraitsType_>
using ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::ABCGridDescs
Initial value:
remove_cvref_t<
decltype(ConvToGemmTransformer{}.MakeABCGridDescriptor_A_K0_M_K1_B_K0_N_K1_C_M_N())>
ck::remove_cvref_t
remove_cv_t< remove_reference_t< T > > remove_cvref_t
Definition type.hpp:297
ck_tile::GroupedConvBwdDataKernelArgs::ConvToGemmTransformer
TransformConvBwdDataToGemm< GroupedConvTraitsType_::NDimSpatial, GroupedConvTraitsType_::ConvSpecialization, GroupedConvTraitsType_::VectorSizeA, GroupedConvTraitsType_::VectorSizeB, GroupedConvTraitsType_::VectorSizeC, true > ConvToGemmTransformer
Definition grouped_convolution_backward_data_kernel.hpp:25

◆ AGridDescKM

template<typename GroupedConvTraitsType_>
using ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::AGridDescKM = remove_cvref_t<decltype(ABCGridDescs{}[number<0>{}])>

◆ BGridDescKN

template<typename GroupedConvTraitsType_>
using ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::BGridDescKN = remove_cvref_t<decltype(ABCGridDescs{}[number<1>{}])>

◆ CGridDescMN

template<typename GroupedConvTraitsType_>
using ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::CGridDescMN = remove_cvref_t<decltype(ABCGridDescs{}[number<2>{}])>

◆ ConvToGemmTransformer

template<typename GroupedConvTraitsType_>
using ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::ConvToGemmTransformer
Initial value:
TransformConvBwdWeightToGemm<GroupedConvTraitsType_::NDimSpatial,
GroupedConvTraitsType_::ConvSpecialization,
GroupedConvTraitsType_::VectorSizeA,
GroupedConvTraitsType_::VectorSizeB,
GroupedConvTraitsType_::VectorSizeC,
GroupedConvTraitsType_::NumGroupsToMerge>
ck_tile::TransformConvBwdWeightToGemm
Definition tile/ops/grouped_convolution/utils/transform_conv_bwd_weight_to_gemm.hpp:22

Constructor & Destructor Documentation

◆ GroupedConvBwdWeightKernelArgs() [1/3]

template<typename GroupedConvTraitsType_>
template<typename InLay = typename GroupedConvTraitsType_::InLayout, typename WeiLay = typename GroupedConvTraitsType_::WeiLayout, typename OutLay = typename GroupedConvTraitsType_::OutLayout, typename std::enable_if< std::is_same_v< InLay, tensor_layout::convolution::NWGC > &&std::is_same_v< WeiLay, tensor_layout::convolution::GKXC > &&std::is_same_v< OutLay, tensor_layout::convolution::NWGK >, bool >::type = false>
CK_TILE_HOST ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::GroupedConvBwdWeightKernelArgs ( const GroupedConvBwdWeightHostArgs & args)
inline

◆ GroupedConvBwdWeightKernelArgs() [2/3]

template<typename GroupedConvTraitsType_>
template<typename InLay = typename GroupedConvTraitsType_::InLayout, typename WeiLay = typename GroupedConvTraitsType_::WeiLayout, typename OutLay = typename GroupedConvTraitsType_::OutLayout, typename std::enable_if< std::is_same_v< InLay, tensor_layout::convolution::NHWGC > &&std::is_same_v< WeiLay, tensor_layout::convolution::GKYXC > &&std::is_same_v< OutLay, tensor_layout::convolution::NHWGK >, bool >::type = false>
CK_TILE_HOST ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::GroupedConvBwdWeightKernelArgs ( const GroupedConvBwdWeightHostArgs & args)
inline

◆ GroupedConvBwdWeightKernelArgs() [3/3]

template<typename GroupedConvTraitsType_>
template<typename InLay = typename GroupedConvTraitsType_::InLayout, typename WeiLay = typename GroupedConvTraitsType_::WeiLayout, typename OutLay = typename GroupedConvTraitsType_::OutLayout, typename std::enable_if< std::is_same_v< InLay, tensor_layout::convolution::NDHWGC > &&std::is_same_v< WeiLay, tensor_layout::convolution::GKZYXC > &&std::is_same_v< OutLay, tensor_layout::convolution::NDHWGK >, bool >::type = false>
CK_TILE_HOST ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::GroupedConvBwdWeightKernelArgs ( const GroupedConvBwdWeightHostArgs & args)
inline

Member Data Documentation

◆ a_grid_desc_k_m

template<typename GroupedConvTraitsType_>
AGridDescKM ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::a_grid_desc_k_m

◆ b_grid_desc_k_n

template<typename GroupedConvTraitsType_>
BGridDescKN ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::b_grid_desc_k_n

◆ c_grid_desc_m_n

template<typename GroupedConvTraitsType_>
CGridDescMN ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::c_grid_desc_m_n

◆ conv_filter_dilations

template<typename GroupedConvTraitsType_>
array<index_t, GroupedConvTraitsType_::NDimSpatial> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::conv_filter_dilations

◆ conv_filter_strides

template<typename GroupedConvTraitsType_>
array<index_t, GroupedConvTraitsType_::NDimSpatial> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::conv_filter_strides

◆ ds_ptr

template<typename GroupedConvTraitsType_>
std::array<const void*, NumDTensor> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::ds_ptr

◆ GemmBatch

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::GemmBatch

◆ GemmK

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::GemmK

◆ GemmM

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::GemmM

◆ GemmN

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::GemmN

◆ group_stride_a

template<typename GroupedConvTraitsType_>
long_index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::group_stride_a

◆ group_stride_b

template<typename GroupedConvTraitsType_>
long_index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::group_stride_b

◆ group_stride_c

template<typename GroupedConvTraitsType_>
long_index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::group_stride_c

◆ in_g_n_c_wis_lengths

template<typename GroupedConvTraitsType_>
array<index_t, NonSpatialDims + GroupedConvTraitsType_::NDimSpatial> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::in_g_n_c_wis_lengths

◆ in_ptr

template<typename GroupedConvTraitsType_>
const void* ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::in_ptr

◆ input_left_pads

template<typename GroupedConvTraitsType_>
array<index_t, GroupedConvTraitsType_::NDimSpatial> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::input_left_pads

◆ input_right_pads

template<typename GroupedConvTraitsType_>
array<index_t, GroupedConvTraitsType_::NDimSpatial> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::input_right_pads

◆ k_batch

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::k_batch

◆ NonSpatialDims

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::NonSpatialDims = 3
staticconstexpr

◆ NumDTensor

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::NumDTensor = GroupedConvTraitsType_::NumDTensor
staticconstexpr

◆ NumGroupsPerBatch

template<typename GroupedConvTraitsType_>
index_t ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::NumGroupsPerBatch

◆ out_g_n_k_wos_lengths

template<typename GroupedConvTraitsType_>
array<index_t, NonSpatialDims + GroupedConvTraitsType_::NDimSpatial> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::out_g_n_k_wos_lengths

◆ out_ptr

template<typename GroupedConvTraitsType_>
const void* ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::out_ptr

◆ wei_g_k_c_xs_lengths

template<typename GroupedConvTraitsType_>
array<index_t, NonSpatialDims + GroupedConvTraitsType_::NDimSpatial> ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::wei_g_k_c_xs_lengths

◆ wei_ptr

template<typename GroupedConvTraitsType_>
void* ck_tile::GroupedConvBwdWeightKernelArgs< GroupedConvTraitsType_ >::wei_ptr
The documentation for this struct was generated from the following file: