cutlass/docs/include_2cutlass_2gemm_2dev...

194 lines
157 KiB
HTML
Raw Permalink Normal View History

<!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: gemm.h Source File</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><a href="annotated.html"><span>Classes</span></a></li>
<li class="current"><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="files.html"><span>File&#160;List</span></a></li>
<li><a href="globals.html"><span>File&#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="dir_d44c64559bbebec7f509842c48db8b23.html">include</a></li><li class="navelem"><a class="el" href="dir_6baf2bb612a2f0daa69af3101ede80a1.html">cutlass</a></li><li class="navelem"><a class="el" href="dir_9aa36bd9cfad59a1f88859a38871c977.html">gemm</a></li><li class="navelem"><a class="el" href="dir_36528dc2736efa40b421028b7309c671.html">device</a></li> </ul>
</div>
</div><!-- top -->
<div class="header">
<div class="headertitle">
<div class="title">include/cutlass/gemm/device/gemm.h</div> </div>
</div><!--header-->
<div class="contents">
<a href="include_2cutlass_2gemm_2device_2gemm_8h.html">Go to the documentation of this file.</a><div class="fragment"><div class="line"><a name="l00001"></a><span class="lineno"> 1</span>&#160;<span class="comment">/***************************************************************************************************</span></div><div class="line"><a name="l00002"></a><span class="lineno"> 2</span>&#160;<span class="comment"> * Copyright (c) 2017-2019, NVIDIA CORPORATION. All rights reserved.</span></div><div class="line"><a name="l00003"></a><span class="lineno"> 3</span>&#160;<span class="comment"> *</span></div><div class="line"><a name="l00004"></a><span class="lineno"> 4</span>&#160;<span class="comment"> * Redistribution and use in source and binary forms, with or without modification, are permitted</span></div><div class="line"><a name="l00005"></a><span class="lineno"> 5</span>&#160;<span class="comment"> * provided that the following conditions are met:</span></div><div class="line"><a name="l00006"></a><span class="lineno"> 6</span>&#160;<span class="comment"> * * Redistributions of source code must retain the above copyright notice, this list of</span></div><div class="line"><a name="l00007"></a><span class="lineno"> 7</span>&#160;<span class="comment"> * conditions and the following disclaimer.</span></div><div class="line"><a name="l00008"></a><span class="lineno"> 8</span>&#160;<span class="comment"> * * Redistributions in binary form must reproduce the above copyright notice, this list of</span></div><div class="line"><a name="l00009"></a><span class="lineno"> 9</span>&#160;<span class="comment"> * conditions and the following disclaimer in the documentation and/or other materials</span></div><div class="line"><a name="l00010"></a><span class="lineno"> 10</span>&#160;<span class="comment"> * provided with the distribution.</span></div><div class="line"><a name="l00011"></a><span class="lineno"> 11</span>&#160;<span class="comment"> * * Neither the name of the NVIDIA CORPORATION nor the names of its contributors may be used</span></div><div class="line"><a name="l00012"></a><span class="lineno"> 12</span>&#160;<span class="comment"> * to endorse or promote products derived from this software without specific prior written</span></div><div class="line"><a name="l00013"></a><span class="lineno"> 13</span>&#160;<span class="comment"> * permission.</span></div><div class="line"><a name="l00014"></a><span class="lineno"> 14</span>&#160;<span class="comment"> *</span></div><div class="line"><a name="l00015"></a><span class="lineno"> 15</span>&#160;<span class="comment"> * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS &quot;AS IS&quot; AND ANY EXPRESS OR</span></div><div class="line"><a name="l00016"></a><span class="lineno"> 16</span>&#160;<span class="comment"> * IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED WARRANTIES OF MERCHANTABILITY AND</span></div><div class="line"><a name="l00017"></a><span class="lineno"> 17</span>&#160;<span class="comment"> * FITNESS FOR A PARTICULAR PURPOSE ARE DISCLAIMED. IN NO EVENT SHALL NVIDIA CORPORATION BE LIABLE</span></div><div class="line"><a name="l00018"></a><span class="lineno"> 18</span>&#160;<span class="comment"> * FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL DAMAGES (INCLUDING,</span></div><div class="line"><a name="l00019"></a><span class="lineno"> 19</span>&#160;<span class="comment"> * BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS;</span></div><div class="line"><a name="l00020"></a><span class="lineno"> 20</span>&#160;<span class="comment"> * OR BUSINESS INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT,</span></div><div class="line"><a name="l00021"></a><span class="lineno"> 21</span>&#160;<span class="comment"> * STRICT LIABILITY, OR TOR (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE</span></div><div class
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a5cd83a90660626f6c446d45881f0fc22"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a5cd83a90660626f6c446d45881f0fc22">cutlass::gemm::device::Gemm::kStages</a></div><div class="ttdeci">static int const kStages</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:238</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_ad68cd06dadc163a13f5ed29e07d6535b"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#ad68cd06dadc163a13f5ed29e07d6535b">cutlass::gemm::device::Gemm::Arguments::problem_size</a></div><div class="ttdeci">GemmCoord problem_size</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:276</div></div>
<div class="ttc" id="namespacecutlass_html"><div class="ttname"><a href="namespacecutlass.html">cutlass</a></div><div class="ttdef"><b>Definition:</b> aligned_buffer.h:35</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_aaef8450711318fa1a53fe3cb72b59263"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#aaef8450711318fa1a53fe3cb72b59263">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::split_k_slices</a></div><div class="ttdeci">int split_k_slices</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:606</div></div>
<div class="ttc" id="namespacecutlass_html_ac5a88c5840a28a9e0206b9cc7812a18da58c16acfe39125423ff22c907487ba2c"><div class="ttname"><a href="namespacecutlass.html#ac5a88c5840a28a9e0206b9cc7812a18da58c16acfe39125423ff22c907487ba2c">cutlass::Status::kErrorInvalidProblem</a></div><div class="ttdoc">Specified problem size is not supported by operator. </div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_adf1ec9ab942ed20321cb91f02154abf0"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#adf1ec9ab942ed20321cb91f02154abf0">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::ElementA</a></div><div class="ttdeci">ElementB ElementA</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:219</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_a9bdaf3563983efcca649460be169b334"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#a9bdaf3563983efcca649460be169b334">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::ref_A</a></div><div class="ttdeci">TensorRef&lt; ElementA const, LayoutA &gt; ref_A</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:601</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a0fec423a58e8de8ff7b015e5167ac614"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a0fec423a58e8de8ff7b015e5167ac614">cutlass::gemm::device::Gemm::get_workspace_size</a></div><div class="ttdeci">static size_t get_workspace_size(Arguments const &amp;args)</div><div class="ttdoc">Gets the workspace size. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:350</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a1eda40e6a86fb3ebeabed2f717e47ced"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a1eda40e6a86fb3ebeabed2f717e47ced">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::ThreadblockSwizzle</a></div><div class="ttdeci">ThreadblockSwizzle ThreadblockSwizzle</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:236</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_ac6c397a181a52c0dbb39bf3710ee4658"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#ac6c397a181a52c0dbb39bf3710ee4658">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::Arguments</a></div><div class="ttdeci">CUTLASS_HOST_DEVICE Arguments()</div><div class="ttdoc">Default ctor. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:614</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a40ad889da7ff420fd9f9000cd9f98e32"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a40ad889da7ff420fd9f9000cd9f98e32">cutlass::gemm::device::Gemm::can_implement</a></div><div class="ttdeci">static Status can_implement(Arguments const &amp;args)</div><div class="ttdoc">Determines whether the GEMM can execute the given problem. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:328</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_a4b5c3e7a0f2307ca90632fa476aac1aa"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#a4b5c3e7a0f2307ca90632fa476aac1aa">cutlass::gemm::device::Gemm::Arguments::Arguments</a></div><div class="ttdeci">CUTLASS_HOST_DEVICE Arguments()</div><div class="ttdoc">Default ctor. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:290</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_a975d1d55ab7d42eb02eb226e2bbb5812"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#a975d1d55ab7d42eb02eb226e2bbb5812">cutlass::gemm::device::Gemm::Arguments::Arguments</a></div><div class="ttdeci">CUTLASS_HOST_DEVICE Arguments(GemmCoord problem_size_, TensorRef&lt; ElementA const, LayoutA &gt; ref_A_, TensorRef&lt; ElementB const, LayoutB &gt; ref_B_, TensorRef&lt; ElementC const, LayoutC &gt; ref_C_, TensorRef&lt; ElementC, LayoutC &gt; ref_D_, typename EpilogueOutputOp::Params epilogue_=typename EpilogueOutputOp::Params(), int split_k_slices=1)</div><div class="ttdoc">Constructs an Arguments structure. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:296</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1GemmCoord_html"><div class="ttname"><a href="structcutlass_1_1gemm_1_1GemmCoord.html">cutlass::gemm::GemmCoord</a></div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/gemm.h:94</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a1469133c30fde6b28296e3ff6951e7a4"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a1469133c30fde6b28296e3ff6951e7a4">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::get_workspace_size</a></div><div class="ttdeci">static size_t get_workspace_size(Arguments const &amp;args)</div><div class="ttdoc">Gets the workspace size. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:666</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html">cutlass::gemm::device::Gemm</a></div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:216</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1GemmCoord_html_a1b29d2cb15360ad5499216859ad5436a"><div class="ttname"><a href="structcutlass_1_1gemm_1_1GemmCoord.html#a1b29d2cb15360ad5499216859ad5436a">cutlass::gemm::GemmCoord::n</a></div><div class="ttdeci">CUTLASS_HOST_DEVICE Index const &amp; n() const </div><div class="ttdoc">Returns the GEMM N coordinate. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/gemm.h:137</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_af3c79a0271e684d93c3dca5ad230f45f"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#af3c79a0271e684d93c3dca5ad230f45f">cutlass::gemm::device::Gemm::kSplitKSerial</a></div><div class="ttdeci">static bool const kSplitKSerial</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:242</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a2b6c5275c173d73cffe8e6b6b1ccf2c1"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a2b6c5275c173d73cffe8e6b6b1ccf2c1">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::update</a></div><div class="ttdeci">Status update(Arguments const &amp;args, void *workspace=nullptr)</div><div class="ttdoc">Lightweight update given a subset of arguments. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:678</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_a98c7946bf72f054b026bb6fd49175e90"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#a98c7946bf72f054b026bb6fd49175e90">cutlass::gemm::device::Gemm::Arguments::epilogue</a></div><div class="ttdeci">EpilogueOutputOp::Params epilogue</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:281</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_a426f402c08be99849a4477a07f010a5e"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#a426f402c08be99849a4477a07f010a5e">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::epilogue</a></div><div class="ttdeci">EpilogueOutputOp::Params epilogue</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:605</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_a390abae1ca0e01a4b6e58f3724b48eed"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#a390abae1ca0e01a4b6e58f3724b48eed">cutlass::gemm::device::Gemm::Arguments::ref_A</a></div><div class="ttdeci">TensorRef&lt; ElementA const, LayoutA &gt; ref_A</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:277</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_abcacf502806db50eb17a6d925aee16d5"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#abcacf502806db50eb17a6d925aee16d5">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Gemm</a></div><div class="ttdeci">Gemm()</div><div class="ttdoc">Constructs the GEMM. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:644</div></div>
<div class="ttc" id="classcutlass_1_1layout_1_1ColumnMajor_html"><div class="ttname"><a href="classcutlass_1_1layout_1_1ColumnMajor.html">cutlass::layout::ColumnMajor</a></div><div class="ttdoc">Mapping function for column-major matrices. </div><div class="ttdef"><b>Definition:</b> layout/matrix.h:142</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a662bcbcb6164c803ab490c86e69b9ee1"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a662bcbcb6164c803ab490c86e69b9ee1">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::can_implement</a></div><div class="ttdeci">static Status can_implement(Arguments const &amp;args)</div><div class="ttdoc">Determines whether the GEMM can execute the given problem. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:660</div></div>
<div class="ttc" id="default__gemm_8h_html"><div class="ttname"><a href="default__gemm_8h.html">default_gemm.h</a></div><div class="ttdoc">Default kernel-level GEMM definitions combine threadblock-scoped matrix multiply-add with the appropr...</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a384db4125183e504fafc5a946b7ba757"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a384db4125183e504fafc5a946b7ba757">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::operator()</a></div><div class="ttdeci">Status operator()(cudaStream_t stream=nullptr)</div><div class="ttdoc">Runs the kernel using initialized state. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:690</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_ac98211b8adf5c18b4b6c54c5d1cdbb1a"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#ac98211b8adf5c18b4b6c54c5d1cdbb1a">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::InstructionShape</a></div><div class="ttdeci">InstructionShape InstructionShape</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:234</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a6f8a0ff6be313d9f2aac2de4259f65b4"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a6f8a0ff6be313d9f2aac2de4259f65b4">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::Operator</a></div><div class="ttdeci">Operator Operator</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:237</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_aaaa871717d2fbe254a434160bc5acc65"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#aaaa871717d2fbe254a434160bc5acc65">cutlass::gemm::device::Gemm::update</a></div><div class="ttdeci">Status update(Arguments const &amp;args, void *workspace=nullptr)</div><div class="ttdoc">Lightweight update given a subset of arguments. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:417</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_ac9c3c764f72e29c3aea99a8f3998e6cd"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#ac9c3c764f72e29c3aea99a8f3998e6cd">cutlass::gemm::device::Gemm::Arguments::split_k_slices</a></div><div class="ttdeci">int split_k_slices</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:282</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_af4c31798e982518db6a8fe66e63962e6"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#af4c31798e982518db6a8fe66e63962e6">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::ElementC</a></div><div class="ttdeci">ElementC ElementC</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:225</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a5f4f93ca97b358b4410f3d0b1e0a6387"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a5f4f93ca97b358b4410f3d0b1e0a6387">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::run</a></div><div class="ttdeci">Status run(cudaStream_t stream=nullptr)</div><div class="ttdoc">Runs the kernel using initialized state. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:684</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a114b9c14e102d333ef3dcad7865a4efb"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a114b9c14e102d333ef3dcad7865a4efb">cutlass::gemm::device::Gemm::operator()</a></div><div class="ttdeci">Status operator()(cudaStream_t stream=nullptr)</div><div class="ttdoc">Runs the kernel using initialized state. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:471</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a359e8eaddc486fcd0e15b879b11247ad"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a359e8eaddc486fcd0e15b879b11247ad">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::OperatorClass</a></div><div class="ttdeci">OperatorClass OperatorClass</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:230</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a7615ad046304360243729c29c65e878a"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a7615ad046304360243729c29c65e878a">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::UnderlyingArguments</a></div><div class="ttdeci">typename UnderlyingOperator::Arguments UnderlyingArguments</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:589</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_ad35f92eaf14c268f8bda2a0e48399a9a"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#ad35f92eaf14c268f8bda2a0e48399a9a">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::ElementAccumulator</a></div><div class="ttdeci">ElementAccumulator ElementAccumulator</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:229</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a431ea9cb851566df881457a0f496b976"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a431ea9cb851566df881457a0f496b976">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::GemmKernel</a></div><div class="ttdeci">typename kernel::DefaultGemm&lt; ElementA, LayoutA, kAlignmentA, ElementB, LayoutB, kAlignmentB, ElementC, LayoutC, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, kStages, kSplitKSerial, Operator, kIsBetaZero &gt;::GemmKernel GemmKernel</div><div class="ttdoc">Define the kernel. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:267</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a30809c4843226401b9d2da5ecb8d838b"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a30809c4843226401b9d2da5ecb8d838b">cutlass::gemm::device::Gemm::kAlignmentB</a></div><div class="ttdeci">static int const kAlignmentB</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:240</div></div>
<div class="ttc" id="structcutlass_1_1layout_1_1LayoutTranspose_html"><div class="ttname"><a href="structcutlass_1_1layout_1_1LayoutTranspose.html">cutlass::layout::LayoutTranspose</a></div><div class="ttdoc">Defines transposes of matrix layouts. </div><div class="ttdef"><b>Definition:</b> layout/matrix.h:921</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_ad0f1a2669ecc62f24922613e9bad857f"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#ad0f1a2669ecc62f24922613e9bad857f">cutlass::gemm::device::Gemm::kAlignmentA</a></div><div class="ttdeci">static int const kAlignmentA</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:239</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a01ef9ea6588f488e8d727e36bdec8ba8"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a01ef9ea6588f488e8d727e36bdec8ba8">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::ThreadblockShape</a></div><div class="ttdeci">ThreadblockShape ThreadblockShape</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:232</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_abdba57a68d6982fffbb1cc3db34ef0f9"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#abdba57a68d6982fffbb1cc3db34ef0f9">cutlass::gemm::device::Gemm::Arguments::ref_C</a></div><div class="ttdeci">TensorRef&lt; ElementC const, LayoutC &gt; ref_C</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:279</div></div>
<div class="ttc" id="classcutlass_1_1TensorRef_html"><div class="ttname"><a href="classcutlass_1_1TensorRef.html">cutlass::TensorRef&lt; ElementA const, LayoutA &gt;</a></div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_ae712c362f83fbd45679a6e989315d3dc"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#ae712c362f83fbd45679a6e989315d3dc">cutlass::gemm::device::Gemm::Arguments::ref_B</a></div><div class="ttdeci">TensorRef&lt; ElementB const, LayoutB &gt; ref_B</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:278</div></div>
<div class="ttc" id="namespacecutlass_html_ac5a88c5840a28a9e0206b9cc7812a18da4d3b5847a0b17037c6b69faf8b1d4d71"><div class="ttname"><a href="namespacecutlass.html#ac5a88c5840a28a9e0206b9cc7812a18da4d3b5847a0b17037c6b69faf8b1d4d71">cutlass::Status::kErrorInternal</a></div><div class="ttdoc">An error within CUTLASS occurred. </div></div>
<div class="ttc" id="device__kernel_8h_html"><div class="ttname"><a href="device__kernel_8h.html">device_kernel.h</a></div><div class="ttdoc">Template for generic CUTLASS kernel. </div></div>
<div class="ttc" id="cutlass_8h_html_a28c2443a142676d3d71effdae1a986b1"><div class="ttname"><a href="cutlass_8h.html#a28c2443a142676d3d71effdae1a986b1">CUTLASS_HOST_DEVICE</a></div><div class="ttdeci">#define CUTLASS_HOST_DEVICE</div><div class="ttdef"><b>Definition:</b> cutlass.h:89</div></div>
<div class="ttc" id="numeric__types_8h_html"><div class="ttname"><a href="numeric__types_8h.html">numeric_types.h</a></div><div class="ttdoc">Top-level include for all CUTLASS numeric types. </div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_acd02e86dfff866eade08415e0043ccc3"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#acd02e86dfff866eade08415e0043ccc3">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::problem_size</a></div><div class="ttdeci">GemmCoord problem_size</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:600</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a35d9f803fdfbbd4608243881c04316dc"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a35d9f803fdfbbd4608243881c04316dc">cutlass::gemm::device::Gemm::operator()</a></div><div class="ttdeci">Status operator()(Arguments const &amp;args, void *workspace=nullptr, cudaStream_t stream=nullptr)</div><div class="ttdoc">Runs the kernel using initialized state. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:476</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_aec04d65c6265eb5f63d703f2dd99cb3f"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#aec04d65c6265eb5f63d703f2dd99cb3f">cutlass::gemm::device::Gemm::Gemm</a></div><div class="ttdeci">Gemm()</div><div class="ttdoc">Constructs the GEMM. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:325</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_a331de1adfdcbea6d0137afe64a4f6f4c"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#a331de1adfdcbea6d0137afe64a4f6f4c">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::Arguments</a></div><div class="ttdeci">CUTLASS_HOST_DEVICE Arguments(GemmCoord problem_size_, TensorRef&lt; ElementA const, LayoutA &gt; ref_A_, TensorRef&lt; ElementB const, LayoutB &gt; ref_B_, TensorRef&lt; ElementC const, LayoutC &gt; ref_C_, TensorRef&lt; ElementC, LayoutC &gt; ref_D_, typename EpilogueOutputOp::Params epilogue_=typename EpilogueOutputOp::Params(), int split_k_slices=1)</div><div class="ttdoc">Constructs an Arguments structure. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:618</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a93acaff17a5916bc11040a9ddf04d9f2"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a93acaff17a5916bc11040a9ddf04d9f2">cutlass::gemm::device::Gemm::LayoutC</a></div><div class="ttdeci">LayoutC_ LayoutC</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:226</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_ab389fd311c5763587315c210f40a11f2"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#ab389fd311c5763587315c210f40a11f2">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::LayoutB</a></div><div class="ttdeci">typename layout::LayoutTranspose&lt; LayoutA &gt;::type LayoutB</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:223</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_a590b8da88ae9350042838451e3e37a22"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#a590b8da88ae9350042838451e3e37a22">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::ref_C</a></div><div class="ttdeci">TensorRef&lt; ElementC const, LayoutC &gt; ref_C</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:603</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html">cutlass::gemm::device::Gemm::Arguments</a></div><div class="ttdoc">Argument structure. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:270</div></div>
<div class="ttc" id="default__gemm__configuration_8h_html"><div class="ttname"><a href="default__gemm__configuration_8h.html">default_gemm_configuration.h</a></div><div class="ttdoc">Definitions for GEMM structures. </div></div>
<div class="ttc" id="classcutlass_1_1layout_1_1RowMajor_html"><div class="ttname"><a href="classcutlass_1_1layout_1_1RowMajor.html">cutlass::layout::RowMajor</a></div><div class="ttdoc">Mapping function for row-major matrices. </div><div class="ttdef"><b>Definition:</b> layout/matrix.h:50</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a7a14474e4238d2fac92ad71c6de087d8"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a7a14474e4238d2fac92ad71c6de087d8">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::initialize</a></div><div class="ttdeci">Status initialize(Arguments const &amp;args, void *workspace=nullptr, cudaStream_t stream=nullptr)</div><div class="ttdoc">Initializes GEMM state from arguments. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:672</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_aef8c133e539ef91efc9dba9012118147"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#aef8c133e539ef91efc9dba9012118147">cutlass::gemm::device::Gemm::run</a></div><div class="ttdeci">Status run(cudaStream_t stream=nullptr)</div><div class="ttdoc">Runs the kernel using initialized state. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:435</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_af34b612740d98d3066b2b14785f9e415"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#af34b612740d98d3066b2b14785f9e415">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::EpilogueOutputOp</a></div><div class="ttdeci">EpilogueOutputOp EpilogueOutputOp</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:235</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a53d79d1b434100da1e466e6378ec43ab"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a53d79d1b434100da1e466e6378ec43ab">cutlass::gemm::device::Gemm::initialize</a></div><div class="ttdeci">Status initialize(Arguments const &amp;args, void *workspace=nullptr, cudaStream_t stream=nullptr)</div><div class="ttdoc">Initializes GEMM state from arguments. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:369</div></div>
<div class="ttc" id="namespacecutlass_html_ac5a88c5840a28a9e0206b9cc7812a18da6467125354474c5cff2f3c920abb7d3f"><div class="ttname"><a href="namespacecutlass.html#ac5a88c5840a28a9e0206b9cc7812a18da6467125354474c5cff2f3c920abb7d3f">cutlass::Status::kErrorWorkspaceNull</a></div><div class="ttdoc">The given workspace is null when it is required to be non-null. </div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_adb5bd27efee1575710e6e4157e32449d"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#adb5bd27efee1575710e6e4157e32449d">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::WarpShape</a></div><div class="ttdeci">WarpShape WarpShape</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:233</div></div>
<div class="ttc" id="namespacecutlass_html_ac5a88c5840a28a9e0206b9cc7812a18da8c632159fa131f09d04f94e3cbcd8782"><div class="ttname"><a href="namespacecutlass.html#ac5a88c5840a28a9e0206b9cc7812a18da8c632159fa131f09d04f94e3cbcd8782">cutlass::Status::kSuccess</a></div><div class="ttdoc">Operation was successful. </div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_ac85244d4a91ea8e8c1c3187fad1004db"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#ac85244d4a91ea8e8c1c3187fad1004db">cutlass::gemm::device::Gemm::kAlignmentC</a></div><div class="ttdeci">static int const kAlignmentC</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:241</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_aa9313915a6129f0c43b43ef3698b3ee4"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#aa9313915a6129f0c43b43ef3698b3ee4">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::to_underlying_arguments</a></div><div class="ttdeci">static UnderlyingArguments to_underlying_arguments(Arguments const &amp;args)</div><div class="ttdoc">Helper to construct a transposed equivalent for the underying GEMM operator. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:647</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1GemmCoord_html_a93515a41db6c4b7e9101067f60d41b8c"><div class="ttname"><a href="structcutlass_1_1gemm_1_1GemmCoord.html#a93515a41db6c4b7e9101067f60d41b8c">cutlass::gemm::GemmCoord::m</a></div><div class="ttdeci">CUTLASS_HOST_DEVICE Index const &amp; m() const </div><div class="ttdoc">Returns the GEMM M coordinate. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/gemm.h:129</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a7d04e3640e81787a28956c75acde997e"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a7d04e3640e81787a28956c75acde997e">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::ElementB</a></div><div class="ttdeci">ElementA ElementB</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:222</div></div>
<div class="ttc" id="gemm_2threadblock_2threadblock__swizzle_8h_html"><div class="ttname"><a href="gemm_2threadblock_2threadblock__swizzle_8h.html">threadblock_swizzle.h</a></div><div class="ttdoc">Implements several possible threadblock-swizzling functions mapping blockIdx to GEMM problems...</div></div>
<div class="ttc" id="arch_8h_html"><div class="ttname"><a href="arch_8h.html">arch.h</a></div><div class="ttdoc">Defines tags for architecture-specific configurations. </div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments_html_a66f9983db4a09ac0d90291c0f8723897"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_1_1Arguments.html#a66f9983db4a09ac0d90291c0f8723897">cutlass::gemm::device::Gemm::Arguments::ref_D</a></div><div class="ttdeci">TensorRef&lt; ElementC, LayoutC &gt; ref_D</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:280</div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_ab1d4d5865786a415f87db1def1b029e7"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#ab1d4d5865786a415f87db1def1b029e7">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::ref_D</a></div><div class="ttdeci">TensorRef&lt; ElementC, LayoutC &gt; ref_D</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:604</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a0303e85432228170eba7dc4b418c86b4"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a0303e85432228170eba7dc4b418c86b4">cutlass::gemm::device::Gemm::kIsBetaZero</a></div><div class="ttdeci">static bool const kIsBetaZero</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:243</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a2bdbc5e737f9bfd1e09a7cfb30e60e29"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a2bdbc5e737f9bfd1e09a7cfb30e60e29">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::GemmKernel</a></div><div class="ttdeci">typename UnderlyingOperator::GemmKernel GemmKernel</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:590</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_afe7ab8c15e83c6cd59b6bcf3fe6e48c0"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#afe7ab8c15e83c6cd59b6bcf3fe6e48c0">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::ArchTag</a></div><div class="ttdeci">ArchTag ArchTag</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:231</div></div>
<div class="ttc" id="cutlass_8h_html"><div class="ttname"><a href="cutlass_8h.html">cutlass.h</a></div><div class="ttdoc">Basic include for CUTLASS. </div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c_html_a6115aa957b3ba8ad9e54b7efeefaacd1"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layout4d0960ae6b1d1bf19e6239dbd002249c.html#a6115aa957b3ba8ad9e54b7efeefaacd1">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::operator()</a></div><div class="ttdeci">Status operator()(Arguments const &amp;args, void *workspace=nullptr, cudaStream_t stream=nullptr)</div><div class="ttdoc">Runs the kernel using initialized state. </div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:695</div></div>
<div class="ttc" id="namespacecutlass_html_ac5a88c5840a28a9e0206b9cc7812a18d"><div class="ttname"><a href="namespacecutlass.html#ac5a88c5840a28a9e0206b9cc7812a18d">cutlass::Status</a></div><div class="ttdeci">Status</div><div class="ttdoc">Status code returned by CUTLASS operations. </div><div class="ttdef"><b>Definition:</b> cutlass.h:39</div></div>
<div class="ttc" id="include_2cutlass_2gemm_2kernel_2gemm_8h_html"><div class="ttname"><a href="include_2cutlass_2gemm_2kernel_2gemm_8h.html">gemm.h</a></div><div class="ttdoc">Template for a pipelined GEMM kernel. Does not compute batching or support split-K. </div></div>
<div class="ttc" id="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709_html_ab77204c1010b17c6643d26a89f41c3d0"><div class="ttname"><a href="structcutlass_1_1gemm_1_1device_1_1Gemm_3_01ElementA___00_01LayoutA___00_01ElementB___00_01Layou1b211cc9c97c022d8fe10f2dd32c8709.html#ab77204c1010b17c6643d26a89f41c3d0">cutlass::gemm::device::Gemm&lt; ElementA_, LayoutA_, ElementB_, LayoutB_, ElementC_, layout::ColumnMajor, ElementAccumulator_, OperatorClass_, ArchTag_, ThreadblockShape_, WarpShape_, InstructionShape_, EpilogueOutputOp_, ThreadblockSwizzle_, Stages, AlignmentA, AlignmentB, SplitKSerial, Operator_, IsBetaZero &gt;::Arguments::ref_B</a></div><div class="ttdeci">TensorRef&lt; ElementB const, LayoutB &gt; ref_B</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:602</div></div>
<div class="ttc" id="classcutlass_1_1gemm_1_1device_1_1Gemm_html_a6bbefc44e1ed6822dc8a4e2edc2b276b"><div class="ttname"><a href="classcutlass_1_1gemm_1_1device_1_1Gemm.html#a6bbefc44e1ed6822dc8a4e2edc2b276b">cutlass::gemm::device::Gemm&lt; ElementB, typename layout::LayoutTranspose&lt; LayoutB &gt;::type, ElementA, typename layout::LayoutTranspose&lt; LayoutA &gt;::type, ElementC, layout::RowMajor, ElementAccumulator, OperatorClass, ArchTag, ThreadblockShape, WarpShape, InstructionShape, EpilogueOutputOp, ThreadblockSwizzle, Stages, kAlignmentB, kAlignmentA, SplitKSerial, Operator, kIsBetaZero &gt;::LayoutA</a></div><div class="ttdeci">typename layout::LayoutTranspose&lt; LayoutB &gt;::type LayoutA</div><div class="ttdef"><b>Definition:</b> include/cutlass/gemm/device/gemm.h:220</div></div>
</div><!-- fragment --></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>