GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ > Struct Template Reference

GroupedConvTraits&lt; NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ &gt; Struct Template Reference#

Composable Kernel: ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ > Struct Template Reference
Classes | Public Types | Static Public Attributes | List of all members
ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ > Struct Template Reference

#include <grouped_convolution_utils.hpp>

Classes

struct  FixedGemmParams

Public Types

using InLayout = InLayout_
using WeiLayout = WeiLayout_
using DsLayout = DsLayout_
using OutLayout = OutLayout_
using AsLayoutFwd = ck_tile::tensor_layout::gemm::RowMajor
using BsLayoutFwd = ck_tile::tensor_layout::gemm::ColumnMajor
using CLayoutFwd = ck_tile::tensor_layout::gemm::RowMajor
using AsLayoutBwdData = ck_tile::tensor_layout::gemm::RowMajor
using BsLayoutBwdData = ck_tile::tensor_layout::gemm::RowMajor
using CLayoutBwdData = ck_tile::tensor_layout::gemm::RowMajor
using AsLayoutBwdWeight = ck_tile::tensor_layout::gemm::ColumnMajor
using BsLayoutBwdWeight = ck_tile::tensor_layout::gemm::RowMajor
using CLayoutBwdWeight = ck_tile::tensor_layout::gemm::RowMajor
template<ck_tile::index_t NumWaveGroups = 1>
using GroupedConvImplicitGemmTraitsFwd
template<ck_tile::index_t NumWaveGroups = 1>
using GroupedConvImplicitGemmTraitsBwdData
template<ck_tile::index_t NumWaveGroups = 1>
using GroupedConvImplicitGemmTraitsBwdWeight
using ImplicitGemmDsLayout = decltype(generate_implicit_gemm_layout())

Static Public Attributes

static constexpr bool EnableSplitImage = EnableSplitImage_
static constexpr index_t NumGroupsToMerge = NumGroupsToMerge_
static constexpr index_t NDimSpatial = NDimSpatial_
static constexpr ConvolutionSpecialization ConvSpecialization = ConvSpecialization_
static constexpr ck_tile::index_t VectorSizeA = VectorSizeA_
static constexpr ck_tile::index_t VectorSizeB = VectorSizeB_
static constexpr ck_tile::index_t VectorSizeC = VectorSizeC_
static constexpr ck_tile::index_t NumDTensor = DsLayout::size()

Member Typedef Documentation

◆ AsLayoutBwdData

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::AsLayoutBwdData = ck_tile::tensor_layout::gemm::RowMajor

◆ AsLayoutBwdWeight

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::AsLayoutBwdWeight = ck_tile::tensor_layout::gemm::ColumnMajor

◆ AsLayoutFwd

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::AsLayoutFwd = ck_tile::tensor_layout::gemm::RowMajor

◆ BsLayoutBwdData

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::BsLayoutBwdData = ck_tile::tensor_layout::gemm::RowMajor

◆ BsLayoutBwdWeight

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::BsLayoutBwdWeight = ck_tile::tensor_layout::gemm::RowMajor

◆ BsLayoutFwd

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::BsLayoutFwd = ck_tile::tensor_layout::gemm::ColumnMajor

◆ CLayoutBwdData

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::CLayoutBwdData = ck_tile::tensor_layout::gemm::RowMajor

◆ CLayoutBwdWeight

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::CLayoutBwdWeight = ck_tile::tensor_layout::gemm::RowMajor

◆ CLayoutFwd

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::CLayoutFwd = ck_tile::tensor_layout::gemm::RowMajor

◆ DsLayout

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::DsLayout = DsLayout_

◆ GroupedConvImplicitGemmTraitsBwdData

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
template<ck_tile::index_t NumWaveGroups = 1>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::GroupedConvImplicitGemmTraitsBwdData
Initial value:
TileGemmTraits<true,
true,
true,
AsLayoutBwdData,
BsLayoutBwdData,
CLayoutBwdData,
NumWaveGroups>
ck_tile::GroupedConvTraits::CLayoutBwdData
ck_tile::tensor_layout::gemm::RowMajor CLayoutBwdData
Definition grouped_convolution_utils.hpp:108
ck_tile::GroupedConvTraits::AsLayoutBwdData
ck_tile::tensor_layout::gemm::RowMajor AsLayoutBwdData
Definition grouped_convolution_utils.hpp:106
ck_tile::GroupedConvTraits::BsLayoutBwdData
ck_tile::tensor_layout::gemm::RowMajor BsLayoutBwdData
Definition grouped_convolution_utils.hpp:107
ck_tile::TileGemmTraits
Definition tile_gemm_traits.hpp:18

◆ GroupedConvImplicitGemmTraitsBwdWeight

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
template<ck_tile::index_t NumWaveGroups = 1>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::GroupedConvImplicitGemmTraitsBwdWeight
Initial value:
TileGemmTraits<true,
true,
true,
AsLayoutBwdWeight,
BsLayoutBwdWeight,
CLayoutBwdWeight,
NumWaveGroups>
ck_tile::GroupedConvTraits::BsLayoutBwdWeight
ck_tile::tensor_layout::gemm::RowMajor BsLayoutBwdWeight
Definition grouped_convolution_utils.hpp:111
ck_tile::GroupedConvTraits::AsLayoutBwdWeight
ck_tile::tensor_layout::gemm::ColumnMajor AsLayoutBwdWeight
Definition grouped_convolution_utils.hpp:110
ck_tile::GroupedConvTraits::CLayoutBwdWeight
ck_tile::tensor_layout::gemm::RowMajor CLayoutBwdWeight
Definition grouped_convolution_utils.hpp:112

◆ GroupedConvImplicitGemmTraitsFwd

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
template<ck_tile::index_t NumWaveGroups = 1>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::GroupedConvImplicitGemmTraitsFwd
Initial value:
TileGemmTraits<true, true, true, AsLayoutFwd, BsLayoutFwd, CLayoutFwd, NumWaveGroups>

◆ ImplicitGemmDsLayout

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::ImplicitGemmDsLayout = decltype(generate_implicit_gemm_layout())

◆ InLayout

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::InLayout = InLayout_

◆ OutLayout

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::OutLayout = OutLayout_

◆ WeiLayout

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
using ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::WeiLayout = WeiLayout_

Member Data Documentation

◆ ConvSpecialization

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
ConvolutionSpecialization ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::ConvSpecialization = ConvSpecialization_
staticconstexpr

◆ EnableSplitImage

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
bool ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::EnableSplitImage = EnableSplitImage_
staticconstexpr

◆ NDimSpatial

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
index_t ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::NDimSpatial = NDimSpatial_
staticconstexpr

◆ NumDTensor

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
ck_tile::index_t ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::NumDTensor = DsLayout::size()
staticconstexpr

◆ NumGroupsToMerge

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
index_t ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::NumGroupsToMerge = NumGroupsToMerge_
staticconstexpr

◆ VectorSizeA

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
ck_tile::index_t ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::VectorSizeA = VectorSizeA_
staticconstexpr

◆ VectorSizeB

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
ck_tile::index_t ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::VectorSizeB = VectorSizeB_
staticconstexpr

◆ VectorSizeC

template<index_t NDimSpatial_, ConvolutionSpecialization ConvSpecialization_, typename InLayout_, typename WeiLayout_, typename DsLayout_, typename OutLayout_, index_t VectorSizeA_ = 1, index_t VectorSizeB_ = 1, index_t VectorSizeC_ = 1, index_t NumGroupsToMerge_ = 1, bool EnableSplitImage_ = false>
ck_tile::index_t ck_tile::GroupedConvTraits< NDimSpatial_, ConvSpecialization_, InLayout_, WeiLayout_, DsLayout_, OutLayout_, VectorSizeA_, VectorSizeB_, VectorSizeC_, NumGroupsToMerge_, EnableSplitImage_ >::VectorSizeC = VectorSizeC_
staticconstexpr
The documentation for this struct was generated from the following file: