Update docs

Author: LeiWang1999

api/tilelang.carver.arch.driver.cuda_driver.html

@@ -886,6 +886,11 @@
 <span class="sig-prename descclassname"><span class="pre">tilelang.carver.arch.driver.cuda_driver.</span></span><span class="sig-name descname"><span class="pre">get_persisting_l2_cache_max_size</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="n"><span class="pre">device_id</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">int</span></span><span class="w"> </span><span class="o"><span class="pre">=</span></span><span class="w"> </span><span class="default_value"><span class="pre">0</span></span></em><span class="sig-paren">)</span> <span class="sig-return"><span class="sig-return-icon">&#x2192;</span> <span class="sig-return-typehint"><span class="pre">int</span></span></span><a class="headerlink" href="#tilelang.carver.arch.driver.cuda_driver.get_persisting_l2_cache_max_size" title="Permalink to this definition">#</a></dt>
 <dd></dd></dl>
 
+<dl class="py function">
+<dt class="sig sig-object py" id="tilelang.carver.arch.driver.cuda_driver.get_registers_per_block">
+<span class="sig-prename descclassname"><span class="pre">tilelang.carver.arch.driver.cuda_driver.</span></span><span class="sig-name descname"><span class="pre">get_registers_per_block</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="n"><span class="pre">device_id</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">int</span></span><span class="w"> </span><span class="o"><span class="pre">=</span></span><span class="w"> </span><span class="default_value"><span class="pre">0</span></span></em><span class="sig-paren">)</span> <span class="sig-return"><span class="sig-return-icon">&#x2192;</span> <span class="sig-return-typehint"><span class="pre">int</span></span></span><a class="headerlink" href="#tilelang.carver.arch.driver.cuda_driver.get_registers_per_block" title="Permalink to this definition">#</a></dt>
+<dd></dd></dl>
+
 <dl class="py function">
 <dt class="sig sig-object py" id="tilelang.carver.arch.driver.cuda_driver.get_shared_memory_per_block">
 <span class="sig-prename descclassname"><span class="pre">tilelang.carver.arch.driver.cuda_driver.</span></span><span class="sig-name descname"><span class="pre">get_shared_memory_per_block</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="n"><span class="pre">device_id</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">int</span></span><span class="w"> </span><span class="o"><span class="pre">=</span></span><span class="w"> </span><span class="default_value"><span class="pre">0</span></span></em>, <em class="sig-param"><span class="n"><span class="pre">format</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">str</span></span><span class="w"> </span><span class="o"><span class="pre">=</span></span><span class="w"> </span><span class="default_value"><span class="pre">'bytes'</span></span></em><span class="sig-paren">)</span> <span class="sig-return"><span class="sig-return-icon">&#x2192;</span> <span class="sig-return-typehint"><span class="pre">Optional</span><span class="p"><span class="pre">[</span></span><span class="pre">int</span><span class="p"><span class="pre">]</span></span></span></span><a class="headerlink" href="#tilelang.carver.arch.driver.cuda_driver.get_shared_memory_per_block" title="Permalink to this definition">#</a></dt>
@@ -1027,6 +1032,7 @@
 <li><a class="reference internal" href="#tilelang.carver.arch.driver.cuda_driver.get_max_dynamic_shared_size_bytes"><code class="docutils literal notranslate"><span class="pre">get_max_dynamic_shared_size_bytes()</span></code></a></li>
 <li><a class="reference internal" href="#tilelang.carver.arch.driver.cuda_driver.get_num_sms"><code class="docutils literal notranslate"><span class="pre">get_num_sms()</span></code></a></li>
 <li><a class="reference internal" href="#tilelang.carver.arch.driver.cuda_driver.get_persisting_l2_cache_max_size"><code class="docutils literal notranslate"><span class="pre">get_persisting_l2_cache_max_size()</span></code></a></li>
+<li><a class="reference internal" href="#tilelang.carver.arch.driver.cuda_driver.get_registers_per_block"><code class="docutils literal notranslate"><span class="pre">get_registers_per_block()</span></code></a></li>
 <li><a class="reference internal" href="#tilelang.carver.arch.driver.cuda_driver.get_shared_memory_per_block"><code class="docutils literal notranslate"><span class="pre">get_shared_memory_per_block()</span></code></a></li>
 </ul>
 </li>

api/tilelang.carver.arch.driver.html

@@ -503,6 +503,7 @@ <h2>Submodules<a class="headerlink" href="#submodules" title="Permalink to this
 <li class="toctree-l2"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_max_dynamic_shared_size_bytes"><code class="docutils literal notranslate"><span class="pre">get_max_dynamic_shared_size_bytes()</span></code></a></li>
 <li class="toctree-l2"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_num_sms"><code class="docutils literal notranslate"><span class="pre">get_num_sms()</span></code></a></li>
 <li class="toctree-l2"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_persisting_l2_cache_max_size"><code class="docutils literal notranslate"><span class="pre">get_persisting_l2_cache_max_size()</span></code></a></li>
+<li class="toctree-l2"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_registers_per_block"><code class="docutils literal notranslate"><span class="pre">get_registers_per_block()</span></code></a></li>
 <li class="toctree-l2"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_shared_memory_per_block"><code class="docutils literal notranslate"><span class="pre">get_shared_memory_per_block()</span></code></a></li>
 </ul>
 </li>

api/tilelang.carver.arch.html

@@ -434,6 +434,7 @@ <h2>Subpackages<a class="headerlink" href="#subpackages" title="Permalink to thi
 <li class="toctree-l4"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_max_dynamic_shared_size_bytes"><code class="docutils literal notranslate"><span class="pre">get_max_dynamic_shared_size_bytes()</span></code></a></li>
 <li class="toctree-l4"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_num_sms"><code class="docutils literal notranslate"><span class="pre">get_num_sms()</span></code></a></li>
 <li class="toctree-l4"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_persisting_l2_cache_max_size"><code class="docutils literal notranslate"><span class="pre">get_persisting_l2_cache_max_size()</span></code></a></li>
+<li class="toctree-l4"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_registers_per_block"><code class="docutils literal notranslate"><span class="pre">get_registers_per_block()</span></code></a></li>
 <li class="toctree-l4"><a class="reference internal" href="tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_shared_memory_per_block"><code class="docutils literal notranslate"><span class="pre">get_shared_memory_per_block()</span></code></a></li>
 </ul>
 </li>

api/tilelang.language.logical.html

@@ -423,7 +423,7 @@
 <p>The language interface for tl programs.</p>
 <dl class="py function">
 <dt class="sig sig-object py" id="tilelang.language.logical.all_of">
-<span class="sig-prename descclassname"><span class="pre">tilelang.language.logical.</span></span><span class="sig-name descname"><span class="pre">all_of</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="pre">buffer:</span> <span class="pre">~typing.Union[&lt;tilelang.language.proxy.TensorProxy</span> <span class="pre">object</span> <span class="pre">at</span> <span class="pre">0x7fa62dcc2a40&gt;,</span> <span class="pre">~tvm.tir.stmt.BufferRegion]</span></em><span class="sig-paren">)</span><a class="headerlink" href="#tilelang.language.logical.all_of" title="Permalink to this definition">#</a></dt>
+<span class="sig-prename descclassname"><span class="pre">tilelang.language.logical.</span></span><span class="sig-name descname"><span class="pre">all_of</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="pre">buffer:</span> <span class="pre">~typing.Union[&lt;tilelang.language.proxy.TensorProxy</span> <span class="pre">object</span> <span class="pre">at</span> <span class="pre">0x7ff57f0bdc00&gt;,</span> <span class="pre">~tvm.tir.stmt.BufferRegion]</span></em><span class="sig-paren">)</span><a class="headerlink" href="#tilelang.language.logical.all_of" title="Permalink to this definition">#</a></dt>
 <dd><p>Check if all elements in the buffer are true.</p>
 <dl class="field-list simple">
 <dt class="field-odd">Parameters<span class="colon">:</span></dt>
@@ -437,7 +437,7 @@
 
 <dl class="py function">
 <dt class="sig sig-object py" id="tilelang.language.logical.any_of">
-<span class="sig-prename descclassname"><span class="pre">tilelang.language.logical.</span></span><span class="sig-name descname"><span class="pre">any_of</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="pre">buffer:</span> <span class="pre">~typing.Union[&lt;tilelang.language.proxy.TensorProxy</span> <span class="pre">object</span> <span class="pre">at</span> <span class="pre">0x7fa62dcc2a40&gt;,</span> <span class="pre">~tvm.tir.stmt.BufferRegion]</span></em><span class="sig-paren">)</span><a class="headerlink" href="#tilelang.language.logical.any_of" title="Permalink to this definition">#</a></dt>
+<span class="sig-prename descclassname"><span class="pre">tilelang.language.logical.</span></span><span class="sig-name descname"><span class="pre">any_of</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="pre">buffer:</span> <span class="pre">~typing.Union[&lt;tilelang.language.proxy.TensorProxy</span> <span class="pre">object</span> <span class="pre">at</span> <span class="pre">0x7ff57f0bdc00&gt;,</span> <span class="pre">~tvm.tir.stmt.BufferRegion]</span></em><span class="sig-paren">)</span><a class="headerlink" href="#tilelang.language.logical.any_of" title="Permalink to this definition">#</a></dt>
 <dd><p>Check if any element in the buffer is true.</p>
 <dl class="field-list simple">
 <dt class="field-odd">Parameters<span class="colon">:</span></dt>

api/tilelang.language.proxy.html

@@ -471,7 +471,7 @@
 <p>Buffer proxy class for constructing tir buffer.</p>
 <dl class="py method">
 <dt class="sig sig-object py" id="tilelang.language.proxy.BufferProxy.from_ptr">
-<span class="sig-name descname"><span class="pre">from_ptr</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="n"><span class="pre">pointer_var</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">Var</span></span></em>, <em class="sig-param"><span class="n"><span class="pre">shape</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">tuple</span><span class="p"><span class="pre">[</span></span><span class="pre">tvm.ir.expr.PrimExpr</span><span class="p"><span class="pre">,</span></span><span class="w"> </span><span class="p"><span class="pre">...</span></span><span class="p"><span class="pre">]</span></span></span></em>, <em class="sig-param"><span class="n"><span class="pre">dtype</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">str</span></span><span class="w"> </span><span class="o"><span class="pre">=</span></span><span class="w"> </span><span class="default_value"><span class="pre">'float32'</span></span></em><span class="sig-paren">)</span> <span class="sig-return"><span class="sig-return-icon">&#x2192;</span> <span class="sig-return-typehint"><span class="pre">&lt;tilelang.language.proxy.BufferProxy</span> <span class="pre">object</span> <span class="pre">at</span> <span class="pre">0x7fa62dcc2a70&gt;</span></span></span><a class="headerlink" href="#tilelang.language.proxy.BufferProxy.from_ptr" title="Permalink to this definition">#</a></dt>
+<span class="sig-name descname"><span class="pre">from_ptr</span></span><span class="sig-paren">(</span><em class="sig-param"><span class="n"><span class="pre">pointer_var</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">Var</span></span></em>, <em class="sig-param"><span class="n"><span class="pre">shape</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">tuple</span><span class="p"><span class="pre">[</span></span><span class="pre">tvm.ir.expr.PrimExpr</span><span class="p"><span class="pre">,</span></span><span class="w"> </span><span class="p"><span class="pre">...</span></span><span class="p"><span class="pre">]</span></span></span></em>, <em class="sig-param"><span class="n"><span class="pre">dtype</span></span><span class="p"><span class="pre">:</span></span><span class="w"> </span><span class="n"><span class="pre">str</span></span><span class="w"> </span><span class="o"><span class="pre">=</span></span><span class="w"> </span><span class="default_value"><span class="pre">'float32'</span></span></em><span class="sig-paren">)</span> <span class="sig-return"><span class="sig-return-icon">&#x2192;</span> <span class="sig-return-typehint"><span class="pre">&lt;tilelang.language.proxy.BufferProxy</span> <span class="pre">object</span> <span class="pre">at</span> <span class="pre">0x7ff57f0bdc90&gt;</span></span></span><a class="headerlink" href="#tilelang.language.proxy.BufferProxy.from_ptr" title="Permalink to this definition">#</a></dt>
 <dd><p>Create a buffer from a pointer, shape, and data type.</p>
 <dl class="field-list simple">
 <dt class="field-odd">Parameters<span class="colon">:</span></dt>

genindex.html

@@ -1473,6 +1473,8 @@ <h2>G</h2>
         <li><a href="api/tilelang.carver.roller.node.html#tilelang.carver.roller.node.PrimFuncNode.get_reduce_inputs_dtype">get_reduce_inputs_dtype() (tilelang.carver.roller.node.PrimFuncNode method)</a>
 </li>
         <li><a href="api/tilelang.carver.analysis.html#tilelang.carver.analysis.get_reduction_blocks">get_reduction_blocks() (in module tilelang.carver.analysis)</a>
+</li>
+        <li><a href="api/tilelang.carver.arch.driver.cuda_driver.html#tilelang.carver.arch.driver.cuda_driver.get_registers_per_block">get_registers_per_block() (in module tilelang.carver.arch.driver.cuda_driver)</a>
 </li>
         <li><a href="api/tilelang.carver.utils.html#tilelang.carver.utils.get_roller_hints_from_func">get_roller_hints_from_func() (in module tilelang.carver.utils)</a>
 </li>