ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type > Struct Template Reference

ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow&lt; SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type &gt; Struct Template Reference#

Composable Kernel: ck::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type > Struct Template Reference
ck::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type > Struct Template Reference

#include <threadwise_tensor_slice_transfer.hpp>

Public Types

using Index = MultiIndex< nDim >
 

Public Member Functions

constexpr __device__ ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow (const Index &src_idx)
 
template<typename SrcSliceOriginIdx , typename DstSliceOriginIdx , typename SrcBuffer , typename DstBuffer >
__device__ void Run (const SrcDesc &, const SrcSliceOriginIdx &, const SrcBuffer &src_buf, const DstDesc &, const DstSliceOriginIdx &, DstBuffer &dst_buf) const
 

Public Attributes

ElementwiseOperation element_op_ {}
 

Static Public Attributes

static constexpr index_t nDim = SliceLengths::Size()
 

Member Typedef Documentation

◆ Index

template<typename SrcData , typename DstData , typename SrcDesc , typename DstDesc , typename ElementwiseOperation , typename SliceLengths , typename DimAccessOrder , index_t DstVectorDim, index_t DstScalarPerVector, uint32_t LowEightRowlaneIdx, uint32_t HighEightRowLaneIdx, bool IntraRowSwizzlePerm, typename enable_if< SrcDesc::IsKnownAtCompileTime() &&DstDesc::IsKnownAtCompileTime(), bool >::type = false>
using ck::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type >::Index = MultiIndex<nDim>

Constructor & Destructor Documentation

◆ ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow()

template<typename SrcData , typename DstData , typename SrcDesc , typename DstDesc , typename ElementwiseOperation , typename SliceLengths , typename DimAccessOrder , index_t DstVectorDim, index_t DstScalarPerVector, uint32_t LowEightRowlaneIdx, uint32_t HighEightRowLaneIdx, bool IntraRowSwizzlePerm, typename enable_if< SrcDesc::IsKnownAtCompileTime() &&DstDesc::IsKnownAtCompileTime(), bool >::type = false>
constexpr __device__ ck::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type >::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow ( const Index src_idx)
inlineconstexpr

Member Function Documentation

◆ Run()

template<typename SrcData , typename DstData , typename SrcDesc , typename DstDesc , typename ElementwiseOperation , typename SliceLengths , typename DimAccessOrder , index_t DstVectorDim, index_t DstScalarPerVector, uint32_t LowEightRowlaneIdx, uint32_t HighEightRowLaneIdx, bool IntraRowSwizzlePerm, typename enable_if< SrcDesc::IsKnownAtCompileTime() &&DstDesc::IsKnownAtCompileTime(), bool >::type = false>
template<typename SrcSliceOriginIdx , typename DstSliceOriginIdx , typename SrcBuffer , typename DstBuffer >
__device__ void ck::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type >::Run ( const SrcDesc &  ,
const SrcSliceOriginIdx &  ,
const SrcBuffer &  src_buf,
const DstDesc &  ,
const DstSliceOriginIdx &  ,
DstBuffer &  dst_buf 
) const
inline

Member Data Documentation

◆ element_op_

template<typename SrcData , typename DstData , typename SrcDesc , typename DstDesc , typename ElementwiseOperation , typename SliceLengths , typename DimAccessOrder , index_t DstVectorDim, index_t DstScalarPerVector, uint32_t LowEightRowlaneIdx, uint32_t HighEightRowLaneIdx, bool IntraRowSwizzlePerm, typename enable_if< SrcDesc::IsKnownAtCompileTime() &&DstDesc::IsKnownAtCompileTime(), bool >::type = false>
ElementwiseOperation ck::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type >::element_op_ {}

◆ nDim

template<typename SrcData , typename DstData , typename SrcDesc , typename DstDesc , typename ElementwiseOperation , typename SliceLengths , typename DimAccessOrder , index_t DstVectorDim, index_t DstScalarPerVector, uint32_t LowEightRowlaneIdx, uint32_t HighEightRowLaneIdx, bool IntraRowSwizzlePerm, typename enable_if< SrcDesc::IsKnownAtCompileTime() &&DstDesc::IsKnownAtCompileTime(), bool >::type = false>
constexpr index_t ck::ThreadwiseTensorSliceTransfer_StaticToStatic_InterRow< SrcData, DstData, SrcDesc, DstDesc, ElementwiseOperation, SliceLengths, DimAccessOrder, DstVectorDim, DstScalarPerVector, LowEightRowlaneIdx, HighEightRowLaneIdx, IntraRowSwizzlePerm, type >::nDim = SliceLengths::Size()
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/thread/threadwise_tensor_slice_transfer.hpp