ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose > Struct Template Reference

ThreadGroupTransferGlobal&lt; SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose &gt; Struct Template Reference#

Composable Kernel: ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose > Struct Template Reference
Public Types | Public Member Functions | Static Public Attributes | List of all members
ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose > Struct Template Reference

#include <thread_group_tensor_slice_transfer_global.hpp>

Public Types

using Index = MultiIndex< nDim >
 
using SrcCoord = decltype(make_tensor_coordinate(SrcDesc{}, Index{}))
 
using DstCoord = decltype(make_tensor_coordinate(DstDesc{}, Index{}))
 

Public Member Functions

__device__ ThreadGroupTransferGlobal (const SrcDesc &src_desc, const DstDesc &dst_desc, const Index &src_block_slice_origin, const Index &dst_block_slice_origin, const ElementwiseOperation &element_op)
 
template<typename GridBufferType >
__device__ void RunRead (const SrcDesc &src_desc, const GridBufferType &grid_buf)
 
template<typename BlockBufferType >
__device__ void RunWrite (const DstDesc &dst_desc, BlockBufferType &dst_buf)
 
__device__ void MoveSrcSliceWindow (const SrcDesc &src_desc, const Index &step)
 

Static Public Attributes

static constexpr auto I0 = Number<0>{}
 
static constexpr auto I1 = Number<1>{}
 
static constexpr auto I2 = Number<2>{}
 
static constexpr auto I3 = Number<3>{}
 
static constexpr auto I4 = Number<4>{}
 
static constexpr auto I5 = Number<5>{}
 
static constexpr auto I6 = Number<6>{}
 
static constexpr index_t nDim = remove_reference_t<SrcDesc>::GetNumOfDimension()
 

Member Typedef Documentation

◆ DstCoord

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
using ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::DstCoord = decltype(make_tensor_coordinate(DstDesc{}, Index{}))

◆ Index

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
using ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::Index = MultiIndex<nDim>

◆ SrcCoord

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
using ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::SrcCoord = decltype(make_tensor_coordinate(SrcDesc{}, Index{}))

Constructor & Destructor Documentation

◆ ThreadGroupTransferGlobal()

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
__device__ ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::ThreadGroupTransferGlobal ( const SrcDesc &  src_desc,
const DstDesc &  dst_desc,
const Index &  src_block_slice_origin,
const Index &  dst_block_slice_origin,
const ElementwiseOperation &  element_op 
)
inline

Member Function Documentation

◆ MoveSrcSliceWindow()

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
__device__ void ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::MoveSrcSliceWindow ( const SrcDesc &  src_desc,
const Index &  step 
)
inline

◆ RunRead()

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
template<typename GridBufferType >
__device__ void ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::RunRead ( const SrcDesc &  src_desc,
const GridBufferType &  grid_buf 
)
inline

◆ RunWrite()

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
template<typename BlockBufferType >
__device__ void ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::RunWrite ( const DstDesc &  dst_desc,
BlockBufferType &  dst_buf 
)
inline

Member Data Documentation

◆ I0

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr auto ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::I0 = Number<0>{}
staticconstexpr

◆ I1

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr auto ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::I1 = Number<1>{}
staticconstexpr

◆ I2

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr auto ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::I2 = Number<2>{}
staticconstexpr

◆ I3

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr auto ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::I3 = Number<3>{}
staticconstexpr

◆ I4

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr auto ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::I4 = Number<4>{}
staticconstexpr

◆ I5

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr auto ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::I5 = Number<5>{}
staticconstexpr

◆ I6

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr auto ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::I6 = Number<6>{}
staticconstexpr

◆ nDim

template<typename SrcDesc , typename DstDesc , typename SrcData , typename DstData , typename ElementwiseOperation , typename NumberOfIterations , typename StepsPerIteration , typename IterationOrder , index_t VectorSize, bool DoTranspose>
constexpr index_t ck::ThreadGroupTransferGlobal< SrcDesc, DstDesc, SrcData, DstData, ElementwiseOperation, NumberOfIterations, StepsPerIteration, IterationOrder, VectorSize, DoTranspose >::nDim = remove_reference_t<SrcDesc>::GetNumOfDimension()
staticconstexpr
The documentation for this struct was generated from the following file: