@@ -519,49 +519,46 @@ <h2>Functions<a class="headerlink" href="#functions" title="Link to this heading
519519< tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_cuda_version " title ="tilelang.contrib.nvcc.get_cuda_version "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> get_cuda_version</ span > </ code > </ a > ([cuda_path])</ p > </ td >
520520< td > < p > Utility function to get cuda version</ p > </ td >
521521</ tr >
522- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.tilelang_callback_cuda_compile " title ="tilelang.contrib.nvcc.tilelang_callback_cuda_compile "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> tilelang_callback_cuda_compile</ span > </ code > </ a > (code, target)</ p > </ td >
523- < td > < p > use nvcc to generate fatbin code for better optimization</ p > </ td >
524- </ tr >
525- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.find_libdevice_path " title ="tilelang.contrib.nvcc.find_libdevice_path "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> find_libdevice_path</ span > </ code > </ a > (arch)</ p > </ td >
522+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.find_libdevice_path " title ="tilelang.contrib.nvcc.find_libdevice_path "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> find_libdevice_path</ span > </ code > </ a > (arch)</ p > </ td >
526523< td > < p > Utility function to find libdevice</ p > </ td >
527524</ tr >
528- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.callback_libdevice_path " title ="tilelang.contrib.nvcc.callback_libdevice_path "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> callback_libdevice_path</ span > </ code > </ a > (arch)</ p > </ td >
525+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.callback_libdevice_path " title ="tilelang.contrib.nvcc.callback_libdevice_path "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> callback_libdevice_path</ span > </ code > </ a > (arch)</ p > </ td >
529526< td > < p > </ p > </ td >
530527</ tr >
531- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_target_compute_version " title ="tilelang.contrib.nvcc.get_target_compute_version "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> get_target_compute_version</ span > </ code > </ a > ([target])</ p > </ td >
528+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_target_compute_version " title ="tilelang.contrib.nvcc.get_target_compute_version "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> get_target_compute_version</ span > </ code > </ a > ([target])</ p > </ td >
532529< td > < p > Utility function to get compute capability of compilation target.</ p > </ td >
533530</ tr >
534- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.parse_compute_version " title ="tilelang.contrib.nvcc.parse_compute_version "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> parse_compute_version</ span > </ code > </ a > (compute_version)</ p > </ td >
531+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.parse_compute_version " title ="tilelang.contrib.nvcc.parse_compute_version "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> parse_compute_version</ span > </ code > </ a > (compute_version)</ p > </ td >
535532< td > < p > Parse compute capability string to divide major and minor version</ p > </ td >
536533</ tr >
537- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_target_arch " title ="tilelang.contrib.nvcc.get_target_arch "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> get_target_arch</ span > </ code > </ a > (compute_version)</ p > </ td >
534+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_target_arch " title ="tilelang.contrib.nvcc.get_target_arch "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> get_target_arch</ span > </ code > </ a > (compute_version)</ p > </ td >
538535< td > < p > </ p > </ td >
539536</ tr >
540- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_fp16 " title ="tilelang.contrib.nvcc.have_fp16 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_fp16</ span > </ code > </ a > (compute_version)</ p > </ td >
537+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_fp16 " title ="tilelang.contrib.nvcc.have_fp16 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_fp16</ span > </ code > </ a > (compute_version)</ p > </ td >
541538< td > < p > Either fp16 support is provided in the compute capability or not</ p > </ td >
542539</ tr >
543- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_int8 " title ="tilelang.contrib.nvcc.have_int8 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_int8</ span > </ code > </ a > (compute_version)</ p > </ td >
540+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_int8 " title ="tilelang.contrib.nvcc.have_int8 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_int8</ span > </ code > </ a > (compute_version)</ p > </ td >
544541< td > < p > Either int8 support is provided in the compute capability or not</ p > </ td >
545542</ tr >
546- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_tensorcore " title ="tilelang.contrib.nvcc.have_tensorcore "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_tensorcore</ span > </ code > </ a > ([compute_version, target])</ p > </ td >
543+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_tensorcore " title ="tilelang.contrib.nvcc.have_tensorcore "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_tensorcore</ span > </ code > </ a > ([compute_version, target])</ p > </ td >
547544< td > < p > Either TensorCore support is provided in the compute capability or not</ p > </ td >
548545</ tr >
549- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_cudagraph " title ="tilelang.contrib.nvcc.have_cudagraph "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_cudagraph</ span > </ code > </ a > ()</ p > </ td >
546+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_cudagraph " title ="tilelang.contrib.nvcc.have_cudagraph "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_cudagraph</ span > </ code > </ a > ()</ p > </ td >
550547< td > < p > Either CUDA Graph support is provided</ p > </ td >
551548</ tr >
552- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_bf16 " title ="tilelang.contrib.nvcc.have_bf16 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_bf16</ span > </ code > </ a > (compute_version)</ p > </ td >
549+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_bf16 " title ="tilelang.contrib.nvcc.have_bf16 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_bf16</ span > </ code > </ a > (compute_version)</ p > </ td >
553550< td > < p > Either bf16 support is provided in the compute capability or not</ p > </ td >
554551</ tr >
555- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_fp8 " title ="tilelang.contrib.nvcc.have_fp8 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_fp8</ span > </ code > </ a > (compute_version)</ p > </ td >
552+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_fp8 " title ="tilelang.contrib.nvcc.have_fp8 "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_fp8</ span > </ code > </ a > (compute_version)</ p > </ td >
556553< td > < p > Whether fp8 support is provided in the specified compute capability or not</ p > </ td >
557554</ tr >
558- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_tma " title ="tilelang.contrib.nvcc.have_tma "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_tma</ span > </ code > </ a > (target)</ p > </ td >
555+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.have_tma " title ="tilelang.contrib.nvcc.have_tma "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> have_tma</ span > </ code > </ a > (target)</ p > </ td >
559556< td > < p > Whether TMA support is provided in the specified compute capability or not</ p > </ td >
560557</ tr >
561- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.is_hopper " title ="tilelang.contrib.nvcc.is_hopper "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> is_hopper</ span > </ code > </ a > (target)</ p > </ td >
558+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.is_hopper " title ="tilelang.contrib.nvcc.is_hopper "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> is_hopper</ span > </ code > </ a > (target)</ p > </ td >
562559< td > < p > </ p > </ td >
563560</ tr >
564- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_nvcc_compiler " title ="tilelang.contrib.nvcc.get_nvcc_compiler "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> get_nvcc_compiler</ span > </ code > </ a > ()</ p > </ td >
561+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_nvcc_compiler " title ="tilelang.contrib.nvcc.get_nvcc_compiler "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> get_nvcc_compiler</ span > </ code > </ a > ()</ p > </ td >
565562< td > < p > Get the path to the nvcc compiler</ p > </ td >
566563</ tr >
567564</ tbody >
@@ -687,12 +684,6 @@ <h2>Module Contents<a class="headerlink" href="#module-contents" title="Link to
687684</ dl >
688685</ dd > </ dl >
689686
690- < dl class ="py function ">
691- < dt class ="sig sig-object py " id ="tilelang.contrib.nvcc.tilelang_callback_cuda_compile ">
692- < span class ="sig-prename descclassname "> < span class ="pre "> tilelang.contrib.nvcc.</ span > </ span > < span class ="sig-name descname "> < span class ="pre "> tilelang_callback_cuda_compile</ span > </ span > < span class ="sig-paren "> (</ span > < em class ="sig-param "> < span class ="n "> < span class ="pre "> code</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> target</ span > </ span > </ em > < span class ="sig-paren "> )</ span > < a class ="headerlink " href ="#tilelang.contrib.nvcc.tilelang_callback_cuda_compile " title ="Link to this definition "> ¶</ a > </ dt >
693- < dd > < p > use nvcc to generate fatbin code for better optimization</ p >
694- </ dd > </ dl >
695-
696687< dl class ="py function ">
697688< dt class ="sig sig-object py " id ="tilelang.contrib.nvcc.find_libdevice_path ">
698689< span class ="sig-prename descclassname "> < span class ="pre "> tilelang.contrib.nvcc.</ span > </ span > < span class ="sig-name descname "> < span class ="pre "> find_libdevice_path</ span > </ span > < span class ="sig-paren "> (</ span > < em class ="sig-param "> < span class ="n "> < span class ="pre "> arch</ span > </ span > </ em > < span class ="sig-paren "> )</ span > < a class ="headerlink " href ="#tilelang.contrib.nvcc.find_libdevice_path " title ="Link to this definition "> ¶</ a > </ dt >
@@ -924,7 +915,6 @@ <h2>Module Contents<a class="headerlink" href="#module-contents" title="Link to
924915< li > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_sass_from_source "> < code class ="docutils literal notranslate "> < span class ="pre "> get_sass_from_source()</ span > </ code > </ a > </ li >
925916< li > < a class ="reference internal " href ="#tilelang.contrib.nvcc.find_cuda_path "> < code class ="docutils literal notranslate "> < span class ="pre "> find_cuda_path()</ span > </ code > </ a > </ li >
926917< li > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_cuda_version "> < code class ="docutils literal notranslate "> < span class ="pre "> get_cuda_version()</ span > </ code > </ a > </ li >
927- < li > < a class ="reference internal " href ="#tilelang.contrib.nvcc.tilelang_callback_cuda_compile "> < code class ="docutils literal notranslate "> < span class ="pre "> tilelang_callback_cuda_compile()</ span > </ code > </ a > </ li >
928918< li > < a class ="reference internal " href ="#tilelang.contrib.nvcc.find_libdevice_path "> < code class ="docutils literal notranslate "> < span class ="pre "> find_libdevice_path()</ span > </ code > </ a > </ li >
929919< li > < a class ="reference internal " href ="#tilelang.contrib.nvcc.callback_libdevice_path "> < code class ="docutils literal notranslate "> < span class ="pre "> callback_libdevice_path()</ span > </ code > </ a > </ li >
930920< li > < a class ="reference internal " href ="#tilelang.contrib.nvcc.get_target_compute_version "> < code class ="docutils literal notranslate "> < span class ="pre "> get_target_compute_version()</ span > </ code > </ a > </ li >
0 commit comments