GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector > Struct Template Reference

GridwisePermute&lt; InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector &gt; Struct Template Reference#

Composable Kernel: ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector > Struct Template Reference
Classes | Public Types | Static Public Member Functions | Static Public Attributes | List of all members
ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector > Struct Template Reference

#include <gridwise_permute.hpp>

Classes

struct  Block2TileMap
 

Public Types

using ThisThreadBlock = ThisThreadBlock< BlockSize >
 
using DefaultBlock2TileMap = Block2TileMap
 

Static Public Member Functions

__host__ static constexpr __device__ auto GetInBlockDesc_NPerBlock_HPerBlock_WPerBlock ()
 
template<typename GridDesc >
__host__ static constexpr __device__ auto GetMergedDesc (const GridDesc &desc)
 
__host__ static constexpr __device__ index_t GetSharedMemoryNumberOfByte ()
 
__host__ static constexpr __device__ auto MakeDefaultBlock2TileMap (const InGridDesc &desc)
 
__host__ static constexpr __device__ bool CheckValidity (const InGridDesc &in_grid_desc, const OutGridDesc &out_grid_desc)
 
template<typename Block2TileMap >
static __device__ void Run (const InGridDesc in_grid_desc, const OutGridDesc out_grid_desc, const InDataType *p_in_global, OutDataType *p_out_global, void *__restrict__ p_shared, const ElementwiseOperation elementwise_op, const Block2TileMap &block_2_tile_map)
 

Static Public Attributes

static constexpr auto I0 = Number<0>{}
 
static constexpr auto I1 = Number<1>{}
 
static constexpr auto I2 = Number<2>{}
 

Member Typedef Documentation

◆ DefaultBlock2TileMap

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
using ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::DefaultBlock2TileMap = Block2TileMap

◆ ThisThreadBlock

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
using ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::ThisThreadBlock = ThisThreadBlock<BlockSize>

Member Function Documentation

◆ CheckValidity()

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
__host__ static constexpr __device__ bool ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::CheckValidity ( const InGridDesc &  in_grid_desc,
const OutGridDesc &  out_grid_desc 
)
inlinestaticconstexpr

◆ GetInBlockDesc_NPerBlock_HPerBlock_WPerBlock()

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
__host__ static constexpr __device__ auto ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::GetInBlockDesc_NPerBlock_HPerBlock_WPerBlock ( )
inlinestaticconstexpr

◆ GetMergedDesc()

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
template<typename GridDesc >
__host__ static constexpr __device__ auto ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::GetMergedDesc ( const GridDesc &  desc)
inlinestaticconstexpr

◆ GetSharedMemoryNumberOfByte()

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
__host__ static constexpr __device__ index_t ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::GetSharedMemoryNumberOfByte ( )
inlinestaticconstexpr

◆ MakeDefaultBlock2TileMap()

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
__host__ static constexpr __device__ auto ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::MakeDefaultBlock2TileMap ( const InGridDesc &  desc)
inlinestaticconstexpr

◆ Run()

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
template<typename Block2TileMap >
static __device__ void ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::Run ( const InGridDesc  in_grid_desc,
const OutGridDesc  out_grid_desc,
const InDataType *  p_in_global,
OutDataType *  p_out_global,
void *__restrict__  p_shared,
const ElementwiseOperation  elementwise_op,
const Block2TileMap &  block_2_tile_map 
)
inlinestatic

Member Data Documentation

◆ I0

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
constexpr auto ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::I0 = Number<0>{}
staticconstexpr

◆ I1

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
constexpr auto ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::I1 = Number<1>{}
staticconstexpr

◆ I2

template<typename InGridDesc , typename OutGridDesc , typename InDataType , typename OutDataType , typename ElementwiseOperation , index_t BlockSize, index_t NPerBlock, index_t HPerBlock, index_t WPerBlock, index_t InBlockLdsExtraW, typename InBlockTransferThreadClusterLengths , typename InBlockTransferThreadClusterArrangeOrder , index_t SrcVectorDim, index_t DstVectorDim, index_t SrcScalarPerVector, index_t DstScalarPerVector>
constexpr auto ck::GridwisePermute< InGridDesc, OutGridDesc, InDataType, OutDataType, ElementwiseOperation, BlockSize, NPerBlock, HPerBlock, WPerBlock, InBlockLdsExtraW, InBlockTransferThreadClusterLengths, InBlockTransferThreadClusterArrangeOrder, SrcVectorDim, DstVectorDim, SrcScalarPerVector, DstScalarPerVector >::I2 = Number<2>{}
staticconstexpr
The documentation for this struct was generated from the following file:
  • /home/docs/checkouts/readthedocs.org/user_builds/advanced-micro-devices-composable-kernel/checkouts/docs-6.4.3/include/ck/tensor_operation/gpu/grid/gridwise_permute.hpp