Back to Cutlass

CUTLASS: regular_tile_iterator_pitch_linear_2dthreadtile.h File Reference

docs/regular__tile__iterator__pitch__linear__2dthreadtile_8h.html

4.4.23.7 KB
Original Source

| | CUTLASS

CUDA Templates for Linear Algebra Subroutines and Solvers |

Classes | Namespaces

regular_tile_iterator_pitch_linear_2dthreadtile.h File Reference

Templates implementing loading of tiles from pitch-linear rank=2 tensors. More...

#include "cutlass/cutlass.h"
#include "cutlass/tensor_ref.h"
#include "cutlass/layout/matrix.h"
#include "cutlass/layout/pitch_linear.h"
#include "regular_tile_iterator.h"

Include dependency graph for regular_tile_iterator_pitch_linear_2dthreadtile.h:

This graph shows which files directly or indirectly include this file:

[Go to the source code of this file.](regular tile iterator pitch linear__2dthreadtile_8h_source.html)

|

Classes

| | class | cutlass::transform::threadblock::RegularTileIterator2dThreadTile< Shape, Element, Layout, AdvanceRank, ThreadMap, Alignment > | | | | class | cutlass::transform::threadblock::RegularTileIterator2dThreadTile< Shape_, Element_, layout::PitchLinear, AdvanceRank, ThreadMap_, Alignment > | | | Regular tile iterator specialized for pitch-linear + 2d thread-tiled threadmapping. More...
| | | | class | cutlass::transform::threadblock::RegularTileIterator2dThreadTile< Shape_, Element_, layout::RowMajorInterleaved< 4 >, AdvanceRank, ThreadMap_, Alignment > | | | Regular tile iterator specialized for interleaved layout + 2d thread-tiled threadmapping. More...
| | | | class | cutlass::transform::threadblock::RegularTileIterator2dThreadTile< Shape_, Element_, layout::ColumnMajorInterleaved< 4 >, AdvanceRank, ThreadMap_, Alignment > | | | Regular tile iterator specialized for interleaved layout + 2d thread-tiled threadmapping. More...
| | |

|

Namespaces

| | | cutlass | | | | | cutlass::transform | | | | | cutlass::transform::threadblock | | |

Detailed Description

This iterator uses masks to guard out-of-bounds accesses and visits the last "residue" tile first, with the objective of minimizing predicate mask updates during steady-state operation.

A precomputed "Params" object minimizes the amount of state that must be stored in registers, and integer addition is used to advance the pointer through memory.


Generated by 1.8.11