#include <threadwise_tensor_slice_transfer_v5r1.hpp>
|
| __device__ constexpr | ThreadwiseTensorSliceTransfer_v5r1 (const SrcDesc &src_desc, const Index &src_slice_origin, const DstDesc &dst_desc, const Index &dst_slice_origin) |
| __device__ void | SetSrcSliceOrigin (const SrcDesc &src_desc, const Index &src_slice_origin_idx) |
| __device__ void | SetDstSliceOrigin (const DstDesc &dst_desc, const Index &dst_slice_origin_idx) |
| template<typename SrcBuffer, typename SrcStepHacks> |
| __device__ void | RunRead (const SrcDesc &src_desc, const SrcBuffer &src_buf, const SrcStepHacks &src_step_hacks) |
| template<typename DstBuffer, typename DstStepHacks> |
| __device__ void | RunWrite (const DstDesc &dst_desc, DstBuffer &dst_buf, const DstStepHacks &dst_step_hacks) |
| template<typename SrcBuffer> |
| __device__ void | RunRead (const SrcDesc &src_desc, const SrcBuffer &src_buf) |
| template<typename DstBuffer> |
| __device__ void | RunWrite (const DstDesc &dst_desc, DstBuffer &dst_buf) |
| __device__ void | MoveSrcSliceWindow (const SrcDesc &src_desc, const Index &src_slice_origin_step_idx) |
| template<typename SrcMoveSliceWindowStepHack> |
| __device__ void | MoveSrcSliceWindow (const SrcDesc &src_desc, const Index &src_slice_origin_step_idx, const SrcMoveSliceWindowStepHack &src_move_slice_window_step_hack) |
| __device__ void | MoveDstSliceWindow (const DstDesc &dst_desc, const Index &dst_slice_origin_step_idx) |
◆ DstCoord
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| using ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::DstCoord = decltype(make_tensor_coordinate(DstDesc{}, Index{})) |
◆ DstCoordStep
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| using ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::DstCoordStep = decltype(make_tensor_coordinate_step(DstDesc{}, Index{})) |
◆ Index
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| using ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::Index = MultiIndex<nDim> |
◆ SrcCoord
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| using ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::SrcCoord = decltype(make_tensor_coordinate(SrcDesc{}, Index{})) |
◆ SrcCoordStep
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| using ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::SrcCoordStep = decltype(make_tensor_coordinate_step(SrcDesc{}, Index{})) |
◆ ThreadwiseTensorSliceTransfer_v5r1()
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| __device__ constexpr ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::ThreadwiseTensorSliceTransfer_v5r1 |
( |
const SrcDesc & | src_desc, |
|
|
const Index & | src_slice_origin, |
|
|
const DstDesc & | dst_desc, |
|
|
const Index & | dst_slice_origin ) |
|
inlineconstexpr |
◆ GetDstCoordinateResetStep()
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| __device__ constexpr auto ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::GetDstCoordinateResetStep |
( |
| ) |
|
|
inlinestaticconstexpr |
◆ GetSrcCoordinateResetStep()
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| __device__ constexpr auto ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::GetSrcCoordinateResetStep |
( |
| ) |
|
|
inlinestaticconstexpr |
◆ MoveDstSliceWindow()
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::MoveDstSliceWindow |
( |
const DstDesc & | dst_desc, |
|
|
const Index & | dst_slice_origin_step_idx ) |
|
inline |
◆ MoveSrcSliceWindow() [1/2]
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::MoveSrcSliceWindow |
( |
const SrcDesc & | src_desc, |
|
|
const Index & | src_slice_origin_step_idx ) |
|
inline |
◆ MoveSrcSliceWindow() [2/2]
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
template<typename SrcMoveSliceWindowStepHack>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::MoveSrcSliceWindow |
( |
const SrcDesc & | src_desc, |
|
|
const Index & | src_slice_origin_step_idx, |
|
|
const SrcMoveSliceWindowStepHack & | src_move_slice_window_step_hack ) |
|
inline |
◆ RunRead() [1/2]
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
template<typename SrcBuffer>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::RunRead |
( |
const SrcDesc & | src_desc, |
|
|
const SrcBuffer & | src_buf ) |
|
inline |
◆ RunRead() [2/2]
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
template<typename SrcBuffer, typename SrcStepHacks>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::RunRead |
( |
const SrcDesc & | src_desc, |
|
|
const SrcBuffer & | src_buf, |
|
|
const SrcStepHacks & | src_step_hacks ) |
|
inline |
◆ RunWrite() [1/2]
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
template<typename DstBuffer>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::RunWrite |
( |
const DstDesc & | dst_desc, |
|
|
DstBuffer & | dst_buf ) |
|
inline |
◆ RunWrite() [2/2]
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
template<typename DstBuffer, typename DstStepHacks>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::RunWrite |
( |
const DstDesc & | dst_desc, |
|
|
DstBuffer & | dst_buf, |
|
|
const DstStepHacks & | dst_step_hacks ) |
|
inline |
◆ SetDstSliceOrigin()
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::SetDstSliceOrigin |
( |
const DstDesc & | dst_desc, |
|
|
const Index & | dst_slice_origin_idx ) |
|
inline |
◆ SetSrcSliceOrigin()
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| __device__ void ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::SetSrcSliceOrigin |
( |
const SrcDesc & | src_desc, |
|
|
const Index & | src_slice_origin_idx ) |
|
inline |
◆ I0
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| auto ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::I0 = Number<0>{} |
|
staticconstexpr |
◆ I1
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| auto ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::I1 = Number<1>{} |
|
staticconstexpr |
◆ nDim
template<typename SliceLengths,
InMemoryDataOperationEnum DstInMemOp, typename SrcData, typename DstData, typename SrcDesc, typename DstDesc, typename SrcDimAccessOrder, typename DstDimAccessOrder, typename SrcVectorTensorLengths, typename DstVectorTensorLengths, typename SrcVectorTensorContiguousDimOrder, typename DstVectorTensorContiguousDimOrder, bool SrcResetCoordinateAfterRun, bool DstResetCoordinateAfterRun>
| index_t ck::ThreadwiseTensorSliceTransfer_v5r1< SliceLengths, DstInMemOp, SrcData, DstData, SrcDesc, DstDesc, SrcDimAccessOrder, DstDimAccessOrder, SrcVectorTensorLengths, DstVectorTensorLengths, SrcVectorTensorContiguousDimOrder, DstVectorTensorContiguousDimOrder, SrcResetCoordinateAfterRun, DstResetCoordinateAfterRun >::nDim = SliceLengths::Size() |
|
staticconstexpr |
The documentation for this struct was generated from the following file: