cutlass/docs/classcutlass_1_1epilogue_1_...

580 lines
39 KiB
HTML

<!DOCTYPE html PUBLIC "-//W3C//DTD XHTML 1.0 Transitional//EN" "http://www.w3.org/TR/xhtml1/DTD/xhtml1-transitional.dtd">
<html xmlns="http://www.w3.org/1999/xhtml">
<head>
<meta http-equiv="Content-Type" content="text/xhtml;charset=UTF-8"/>
<meta http-equiv="X-UA-Compatible" content="IE=9"/>
<meta name="generator" content="Doxygen 1.8.11"/>
<title>CUTLASS: cutlass::epilogue::threadblock::SharedLoadIterator&lt; ThreadMap_, Element_, MaxAlignment &gt; Class Template Reference</title>
<link href="tabs.css" rel="stylesheet" type="text/css"/>
<script type="text/javascript" src="jquery.js"></script>
<script type="text/javascript" src="dynsections.js"></script>
<link href="search/search.css" rel="stylesheet" type="text/css"/>
<script type="text/javascript" src="search/searchdata.js"></script>
<script type="text/javascript" src="search/search.js"></script>
<script type="text/javascript">
$(document).ready(function() { init_search(); });
</script>
<script type="text/x-mathjax-config">
MathJax.Hub.Config({
extensions: ["tex2jax.js"],
jax: ["input/TeX","output/HTML-CSS"],
});
</script><script type="text/javascript" src="http://cdn.mathjax.org/mathjax/latest/MathJax.js"></script>
<link href="doxygen.css" rel="stylesheet" type="text/css" />
</head>
<body>
<div id="top"><!-- do not remove this div, it is closed by doxygen! -->
<div id="titlearea">
<table cellspacing="0" cellpadding="0">
<tbody>
<tr style="height: 56px;">
<td id="projectlogo"><img alt="Logo" src="cutlass-logo-small.png"/></td>
<td id="projectalign" style="padding-left: 0.5em;">
<div id="projectname">CUTLASS
</div>
<div id="projectbrief">CUDA Templates for Linear Algebra Subroutines and Solvers</div>
</td>
</tr>
</tbody>
</table>
</div>
<!-- end header part -->
<!-- Generated by Doxygen 1.8.11 -->
<script type="text/javascript">
var searchBox = new SearchBox("searchBox", "search",false,'Search');
</script>
<div id="navrow1" class="tabs">
<ul class="tablist">
<li><a href="index.html"><span>Main&#160;Page</span></a></li>
<li><a href="modules.html"><span>Modules</span></a></li>
<li><a href="namespaces.html"><span>Namespaces</span></a></li>
<li class="current"><a href="annotated.html"><span>Classes</span></a></li>
<li><a href="files.html"><span>Files</span></a></li>
<li>
<div id="MSearchBox" class="MSearchBoxInactive">
<span class="left">
<img id="MSearchSelect" src="search/mag_sel.png"
onmouseover="return searchBox.OnSearchSelectShow()"
onmouseout="return searchBox.OnSearchSelectHide()"
alt=""/>
<input type="text" id="MSearchField" value="Search" accesskey="S"
onfocus="searchBox.OnSearchFieldFocus(true)"
onblur="searchBox.OnSearchFieldFocus(false)"
onkeyup="searchBox.OnSearchFieldChange(event)"/>
</span><span class="right">
<a id="MSearchClose" href="javascript:searchBox.CloseResultsWindow()"><img id="MSearchCloseImg" border="0" src="search/close.png" alt=""/></a>
</span>
</div>
</li>
</ul>
</div>
<div id="navrow2" class="tabs2">
<ul class="tablist">
<li><a href="annotated.html"><span>Class&#160;List</span></a></li>
<li><a href="classes.html"><span>Class&#160;Index</span></a></li>
<li><a href="inherits.html"><span>Class&#160;Hierarchy</span></a></li>
<li><a href="functions.html"><span>Class&#160;Members</span></a></li>
</ul>
</div>
<!-- window showing the filter options -->
<div id="MSearchSelectWindow"
onmouseover="return searchBox.OnSearchSelectShow()"
onmouseout="return searchBox.OnSearchSelectHide()"
onkeydown="return searchBox.OnSearchSelectKey(event)">
</div>
<!-- iframe showing the search results (closed by default) -->
<div id="MSearchResultsWindow">
<iframe src="javascript:void(0)" frameborder="0"
name="MSearchResults" id="MSearchResults">
</iframe>
</div>
<div id="nav-path" class="navpath">
<ul>
<li class="navelem"><a class="el" href="namespacecutlass.html">cutlass</a></li><li class="navelem"><a class="el" href="namespacecutlass_1_1epilogue.html">epilogue</a></li><li class="navelem"><a class="el" href="namespacecutlass_1_1epilogue_1_1threadblock.html">threadblock</a></li><li class="navelem"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">SharedLoadIterator</a></li> </ul>
</div>
</div><!-- top -->
<div class="header">
<div class="summary">
<a href="#pub-types">Public Types</a> &#124;
<a href="#pub-methods">Public Member Functions</a> &#124;
<a href="#pub-static-attribs">Static Public Attributes</a> &#124;
<a href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator-members.html">List of all members</a> </div>
<div class="headertitle">
<div class="title">cutlass::epilogue::threadblock::SharedLoadIterator&lt; ThreadMap_, Element_, MaxAlignment &gt; Class Template Reference</div> </div>
</div><!--header-->
<div class="contents">
<p><code>#include &lt;<a class="el" href="shared__load__iterator_8h_source.html">shared_load_iterator.h</a>&gt;</code></p>
<table class="memberdecls">
<tr class="heading"><td colspan="2"><h2 class="groupheader"><a name="pub-types"></a>
Public Types</h2></td></tr>
<tr class="memitem:a55f1d151cdd69269ac86c3567d1eff28"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a55f1d151cdd69269ac86c3567d1eff28">ThreadMap</a> = ThreadMap_</td></tr>
<tr class="separator:a55f1d151cdd69269ac86c3567d1eff28"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a996bdfe3c2ef7f1d3684ce9a71f3b31d"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a996bdfe3c2ef7f1d3684ce9a71f3b31d">Shape</a> = typename ThreadMap::Shape</td></tr>
<tr class="separator:a996bdfe3c2ef7f1d3684ce9a71f3b31d"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a961521dfce056013892c64a0ecd4ef2f"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a> = Element_</td></tr>
<tr class="separator:a961521dfce056013892c64a0ecd4ef2f"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:abc1ac65c12315e325e02c5c2e313e0a7"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abc1ac65c12315e325e02c5c2e313e0a7">Layout</a> = <a class="el" href="classcutlass_1_1layout_1_1RowMajor.html">layout::RowMajor</a></td></tr>
<tr class="separator:abc1ac65c12315e325e02c5c2e313e0a7"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:ae7ddd9a25483dcd3efb1d314bea81137"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae7ddd9a25483dcd3efb1d314bea81137">TensorRef</a> = <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae7ddd9a25483dcd3efb1d314bea81137">TensorRef</a>&lt; <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a>, <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abc1ac65c12315e325e02c5c2e313e0a7">Layout</a> &gt;</td></tr>
<tr class="separator:ae7ddd9a25483dcd3efb1d314bea81137"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a5208835793dcd89f36ea65a8fcadf7e7"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a5208835793dcd89f36ea65a8fcadf7e7">ConstTensorRef</a> = typename <a class="el" href="classcutlass_1_1TensorRef.html#ad3c5c9466713f62a5191e720827f34da">TensorRef::ConstTensorRef</a></td></tr>
<tr class="separator:a5208835793dcd89f36ea65a8fcadf7e7"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a82ecb0e54bbcb04364a1924167cbeb15"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a82ecb0e54bbcb04364a1924167cbeb15">Index</a> = typename <a class="el" href="classcutlass_1_1layout_1_1RowMajor.html#aa49e242b14b4f482bc6bdd082acfb576">Layout::Index</a></td></tr>
<tr class="separator:a82ecb0e54bbcb04364a1924167cbeb15"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:ae5daed317dfa5a63b781f4900454bcd3"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae5daed317dfa5a63b781f4900454bcd3">LongIndex</a> = typename <a class="el" href="classcutlass_1_1layout_1_1RowMajor.html#a4062a36ab044fdea058504ed52ee60b8">Layout::LongIndex</a></td></tr>
<tr class="separator:ae5daed317dfa5a63b781f4900454bcd3"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:aa30154b84e1eb2d473d389b3e1c56f29"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#aa30154b84e1eb2d473d389b3e1c56f29">TensorCoord</a> = <a class="el" href="structcutlass_1_1MatrixCoord.html">MatrixCoord</a></td></tr>
<tr class="separator:aa30154b84e1eb2d473d389b3e1c56f29"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:afb37ed9a0f3873600b9e743b2dcb805e"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#afb37ed9a0f3873600b9e743b2dcb805e">Fragment</a> = Array&lt; <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a>, ThreadMap::Iterations::kColumn *ThreadMap::Iterations::kRow *ThreadMap::Iterations::kGroup *ThreadMap::Iterations::kCluster *ThreadMap::kElementsPerAccess &gt;</td></tr>
<tr class="memdesc:afb37ed9a0f3873600b9e743b2dcb805e"><td class="mdescLeft">&#160;</td><td class="mdescRight">Fragment object. <a href="#afb37ed9a0f3873600b9e743b2dcb805e">More...</a><br /></td></tr>
<tr class="separator:afb37ed9a0f3873600b9e743b2dcb805e"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a813f447702209b6efdb1d4dea64efc2c"><td class="memItemLeft" align="right" valign="top">using&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a813f447702209b6efdb1d4dea64efc2c">AccessType</a> = <a class="el" href="classcutlass_1_1AlignedArray.html">AlignedArray</a>&lt; <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a>, ThreadMap::kElementsPerAccess, <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae46b49371732c012f0b44399f39d8b0a">kAlignment</a> &gt;</td></tr>
<tr class="memdesc:a813f447702209b6efdb1d4dea64efc2c"><td class="mdescLeft">&#160;</td><td class="mdescRight">Memory access size. <a href="#a813f447702209b6efdb1d4dea64efc2c">More...</a><br /></td></tr>
<tr class="separator:a813f447702209b6efdb1d4dea64efc2c"><td class="memSeparator" colspan="2">&#160;</td></tr>
</table><table class="memberdecls">
<tr class="heading"><td colspan="2"><h2 class="groupheader"><a name="pub-methods"></a>
Public Member Functions</h2></td></tr>
<tr class="memitem:ae031b4f21a36638f51091ad12d529d5a"><td class="memItemLeft" align="right" valign="top">CUTLASS_DEVICE&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae031b4f21a36638f51091ad12d529d5a">SharedLoadIterator</a> (<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae7ddd9a25483dcd3efb1d314bea81137">TensorRef</a> ref, int thread_idx)</td></tr>
<tr class="memdesc:ae031b4f21a36638f51091ad12d529d5a"><td class="mdescLeft">&#160;</td><td class="mdescRight">Constructor. <a href="#ae031b4f21a36638f51091ad12d529d5a">More...</a><br /></td></tr>
<tr class="separator:ae031b4f21a36638f51091ad12d529d5a"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a1e022ec8d968520bfd0bc7ce51dcaf93"><td class="memItemLeft" align="right" valign="top"><a class="el" href="cutlass_8h.html#a28c2443a142676d3d71effdae1a986b1">CUTLASS_HOST_DEVICE</a> void&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a1e022ec8d968520bfd0bc7ce51dcaf93">add_pointer_offset</a> (<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae5daed317dfa5a63b781f4900454bcd3">LongIndex</a> pointer_offset)</td></tr>
<tr class="memdesc:a1e022ec8d968520bfd0bc7ce51dcaf93"><td class="mdescLeft">&#160;</td><td class="mdescRight">Adds a pointer offset in units of Element. <a href="#a1e022ec8d968520bfd0bc7ce51dcaf93">More...</a><br /></td></tr>
<tr class="separator:a1e022ec8d968520bfd0bc7ce51dcaf93"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a4706618cd7b040c01d23440574696417"><td class="memItemLeft" align="right" valign="top">CUTLASS_DEVICE void&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a4706618cd7b040c01d23440574696417">add_tile_offset</a> (<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#aa30154b84e1eb2d473d389b3e1c56f29">TensorCoord</a> const &amp;offset)</td></tr>
<tr class="separator:a4706618cd7b040c01d23440574696417"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a54ab974e2fbeb22d335d55618b3c7b3f"><td class="memItemLeft" align="right" valign="top">CUTLASS_DEVICE void&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a54ab974e2fbeb22d335d55618b3c7b3f">load_with_pointer_offset</a> (<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#afb37ed9a0f3873600b9e743b2dcb805e">Fragment</a> &amp;frag, <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a82ecb0e54bbcb04364a1924167cbeb15">Index</a> pointer_offset)</td></tr>
<tr class="memdesc:a54ab974e2fbeb22d335d55618b3c7b3f"><td class="mdescLeft">&#160;</td><td class="mdescRight">Loads a fragment from memory. <a href="#a54ab974e2fbeb22d335d55618b3c7b3f">More...</a><br /></td></tr>
<tr class="separator:a54ab974e2fbeb22d335d55618b3c7b3f"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:acc9440f6242ee064cb6e48f9387d73ad"><td class="memItemLeft" align="right" valign="top">CUTLASS_DEVICE void&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#acc9440f6242ee064cb6e48f9387d73ad">load</a> (<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#afb37ed9a0f3873600b9e743b2dcb805e">Fragment</a> &amp;frag)</td></tr>
<tr class="memdesc:acc9440f6242ee064cb6e48f9387d73ad"><td class="mdescLeft">&#160;</td><td class="mdescRight">Loads a fragment. <a href="#acc9440f6242ee064cb6e48f9387d73ad">More...</a><br /></td></tr>
<tr class="separator:acc9440f6242ee064cb6e48f9387d73ad"><td class="memSeparator" colspan="2">&#160;</td></tr>
</table><table class="memberdecls">
<tr class="heading"><td colspan="2"><h2 class="groupheader"><a name="pub-static-attribs"></a>
Static Public Attributes</h2></td></tr>
<tr class="memitem:ae7cc22c50e7e00742c7430080d673dc8"><td class="memItemLeft" align="right" valign="top">static int const&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae7cc22c50e7e00742c7430080d673dc8">kElementsPerAccess</a> = ThreadMap::kElementsPerAccess</td></tr>
<tr class="separator:ae7cc22c50e7e00742c7430080d673dc8"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:abd70e40965a6296adeeeb7ddcca7da90"><td class="memItemLeft" align="right" valign="top">static int const&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abd70e40965a6296adeeeb7ddcca7da90">kMinAlignment</a> = ThreadMap_::kElementsPerAccess * <a class="el" href="structcutlass_1_1sizeof__bits.html">sizeof_bits</a>&lt;Element_&gt;::value / 8</td></tr>
<tr class="separator:abd70e40965a6296adeeeb7ddcca7da90"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:ae46b49371732c012f0b44399f39d8b0a"><td class="memItemLeft" align="right" valign="top">static int const&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae46b49371732c012f0b44399f39d8b0a">kAlignment</a> = (MaxAlignment &lt; <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abd70e40965a6296adeeeb7ddcca7da90">kMinAlignment</a> ? MaxAlignment : <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abd70e40965a6296adeeeb7ddcca7da90">kMinAlignment</a>)</td></tr>
<tr class="separator:ae46b49371732c012f0b44399f39d8b0a"><td class="memSeparator" colspan="2">&#160;</td></tr>
<tr class="memitem:a5adc9c1897c49421e3b31fc3f8d34cbc"><td class="memItemLeft" align="right" valign="top">static int const&#160;</td><td class="memItemRight" valign="bottom"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a5adc9c1897c49421e3b31fc3f8d34cbc">kThreads</a> = ThreadMap::kThreads</td></tr>
<tr class="separator:a5adc9c1897c49421e3b31fc3f8d34cbc"><td class="memSeparator" colspan="2">&#160;</td></tr>
</table>
<a name="details" id="details"></a><h2 class="groupheader">Detailed Description</h2>
<div class="textblock"><h3>template&lt;typename ThreadMap_, typename Element_, int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt;<br />
class cutlass::epilogue::threadblock::SharedLoadIterator&lt; ThreadMap_, Element_, MaxAlignment &gt;</h3>
<p>Tile iterator used to load output tile from shared memory in epilogue.</p>
<p>Satisfies: ReadableTileIterator </p>
</div><h2 class="groupheader">Member Typedef Documentation</h2>
<a class="anchor" id="a813f447702209b6efdb1d4dea64efc2c"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a813f447702209b6efdb1d4dea64efc2c">AccessType</a> = <a class="el" href="classcutlass_1_1AlignedArray.html">AlignedArray</a>&lt; <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a>, ThreadMap::kElementsPerAccess, <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae46b49371732c012f0b44399f39d8b0a">kAlignment</a>&gt;</td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a5208835793dcd89f36ea65a8fcadf7e7"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a5208835793dcd89f36ea65a8fcadf7e7">ConstTensorRef</a> = typename <a class="el" href="classcutlass_1_1TensorRef.html#ad3c5c9466713f62a5191e720827f34da">TensorRef::ConstTensorRef</a></td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a961521dfce056013892c64a0ecd4ef2f"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a> = Element_</td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="afb37ed9a0f3873600b9e743b2dcb805e"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#afb37ed9a0f3873600b9e743b2dcb805e">Fragment</a> = Array&lt; <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a>, ThreadMap::Iterations::kColumn * ThreadMap::Iterations::kRow * ThreadMap::Iterations::kGroup * ThreadMap::Iterations::kCluster * ThreadMap::kElementsPerAccess&gt;</td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a82ecb0e54bbcb04364a1924167cbeb15"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a82ecb0e54bbcb04364a1924167cbeb15">Index</a> = typename <a class="el" href="classcutlass_1_1layout_1_1RowMajor.html#aa49e242b14b4f482bc6bdd082acfb576">Layout::Index</a></td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="abc1ac65c12315e325e02c5c2e313e0a7"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abc1ac65c12315e325e02c5c2e313e0a7">Layout</a> = <a class="el" href="classcutlass_1_1layout_1_1RowMajor.html">layout::RowMajor</a></td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="ae5daed317dfa5a63b781f4900454bcd3"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae5daed317dfa5a63b781f4900454bcd3">LongIndex</a> = typename <a class="el" href="classcutlass_1_1layout_1_1RowMajor.html#a4062a36ab044fdea058504ed52ee60b8">Layout::LongIndex</a></td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a996bdfe3c2ef7f1d3684ce9a71f3b31d"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a996bdfe3c2ef7f1d3684ce9a71f3b31d">Shape</a> = typename ThreadMap::Shape</td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="aa30154b84e1eb2d473d389b3e1c56f29"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#aa30154b84e1eb2d473d389b3e1c56f29">TensorCoord</a> = <a class="el" href="structcutlass_1_1MatrixCoord.html">MatrixCoord</a></td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="ae7ddd9a25483dcd3efb1d314bea81137"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae7ddd9a25483dcd3efb1d314bea81137">TensorRef</a> = <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae7ddd9a25483dcd3efb1d314bea81137">TensorRef</a>&lt;<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a961521dfce056013892c64a0ecd4ef2f">Element</a>, <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abc1ac65c12315e325e02c5c2e313e0a7">Layout</a>&gt;</td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a55f1d151cdd69269ac86c3567d1eff28"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="memname">
<tr>
<td class="memname">using <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a55f1d151cdd69269ac86c3567d1eff28">ThreadMap</a> = ThreadMap_</td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<h2 class="groupheader">Constructor &amp; Destructor Documentation</h2>
<a class="anchor" id="ae031b4f21a36638f51091ad12d529d5a"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">CUTLASS_DEVICE <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::<a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">SharedLoadIterator</a> </td>
<td>(</td>
<td class="paramtype"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae7ddd9a25483dcd3efb1d314bea81137">TensorRef</a>&#160;</td>
<td class="paramname"><em>ref</em>, </td>
</tr>
<tr>
<td class="paramkey"></td>
<td></td>
<td class="paramtype">int&#160;</td>
<td class="paramname"><em>thread_idx</em>&#160;</td>
</tr>
<tr>
<td></td>
<td>)</td>
<td></td><td></td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">inline</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<h2 class="groupheader">Member Function Documentation</h2>
<a class="anchor" id="a1e022ec8d968520bfd0bc7ce51dcaf93"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname"><a class="el" href="cutlass_8h.html#a28c2443a142676d3d71effdae1a986b1">CUTLASS_HOST_DEVICE</a> void <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::add_pointer_offset </td>
<td>(</td>
<td class="paramtype"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#ae5daed317dfa5a63b781f4900454bcd3">LongIndex</a>&#160;</td>
<td class="paramname"><em>pointer_offset</em></td><td>)</td>
<td></td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">inline</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a4706618cd7b040c01d23440574696417"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">CUTLASS_DEVICE void <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::add_tile_offset </td>
<td>(</td>
<td class="paramtype"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#aa30154b84e1eb2d473d389b3e1c56f29">TensorCoord</a> const &amp;&#160;</td>
<td class="paramname"><em>offset</em></td><td>)</td>
<td></td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">inline</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="acc9440f6242ee064cb6e48f9387d73ad"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">CUTLASS_DEVICE void <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::load </td>
<td>(</td>
<td class="paramtype"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#afb37ed9a0f3873600b9e743b2dcb805e">Fragment</a> &amp;&#160;</td>
<td class="paramname"><em>frag</em></td><td>)</td>
<td></td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">inline</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a54ab974e2fbeb22d335d55618b3c7b3f"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">CUTLASS_DEVICE void <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::load_with_pointer_offset </td>
<td>(</td>
<td class="paramtype"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#afb37ed9a0f3873600b9e743b2dcb805e">Fragment</a> &amp;&#160;</td>
<td class="paramname"><em>frag</em>, </td>
</tr>
<tr>
<td class="paramkey"></td>
<td></td>
<td class="paramtype"><a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#a82ecb0e54bbcb04364a1924167cbeb15">Index</a>&#160;</td>
<td class="paramname"><em>pointer_offset</em>&#160;</td>
</tr>
<tr>
<td></td>
<td>)</td>
<td></td><td></td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">inline</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<h2 class="groupheader">Member Data Documentation</h2>
<a class="anchor" id="ae46b49371732c012f0b44399f39d8b0a"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">int const <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::kAlignment = (MaxAlignment &lt; <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abd70e40965a6296adeeeb7ddcca7da90">kMinAlignment</a> ? MaxAlignment : <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html#abd70e40965a6296adeeeb7ddcca7da90">kMinAlignment</a>)</td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">static</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="ae7cc22c50e7e00742c7430080d673dc8"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">int const <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::kElementsPerAccess = ThreadMap::kElementsPerAccess</td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">static</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="abd70e40965a6296adeeeb7ddcca7da90"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">int const <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::kMinAlignment = ThreadMap_::kElementsPerAccess * <a class="el" href="structcutlass_1_1sizeof__bits.html">sizeof_bits</a>&lt;Element_&gt;::value / 8</td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">static</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<a class="anchor" id="a5adc9c1897c49421e3b31fc3f8d34cbc"></a>
<div class="memitem">
<div class="memproto">
<div class="memtemplate">
template&lt;typename ThreadMap_ , typename Element_ , int MaxAlignment = ThreadMap_::kElementsPerAccess * sizeof_bits&lt;Element_&gt;::value / 8&gt; </div>
<table class="mlabels">
<tr>
<td class="mlabels-left">
<table class="memname">
<tr>
<td class="memname">int const <a class="el" href="classcutlass_1_1epilogue_1_1threadblock_1_1SharedLoadIterator.html">cutlass::epilogue::threadblock::SharedLoadIterator</a>&lt; ThreadMap_, Element_, MaxAlignment &gt;::kThreads = ThreadMap::kThreads</td>
</tr>
</table>
</td>
<td class="mlabels-right">
<span class="mlabels"><span class="mlabel">static</span></span> </td>
</tr>
</table>
</div><div class="memdoc">
</div>
</div>
<hr/>The documentation for this class was generated from the following file:<ul>
<li><a class="el" href="shared__load__iterator_8h_source.html">shared_load_iterator.h</a></li>
</ul>
</div><!-- contents -->
<!-- start footer part -->
<hr class="footer"/><address class="footer"><small>
Generated by &#160;<a href="http://www.doxygen.org/index.html">
<img class="footer" src="doxygen.png" alt="doxygen"/>
</a> 1.8.11
</small></address>
</body>
</html>