@@ -502,13 +502,22 @@ <h2>Functions<a class="headerlink" href="#functions" title="Link to this heading
502502< tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_absmax " title ="tilelang.language.reduce.reduce_absmax "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> reduce_absmax</ span > </ code > </ a > (buffer, out[, dim, clear])</ p > </ td >
503503< td > < p > Perform reduce absolute max on input buffer, store the result to output buffer.</ p > </ td >
504504</ tr >
505- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.cumsum_fragment " title ="tilelang.language.reduce.cumsum_fragment "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> cumsum_fragment</ span > </ code > </ a > (src, dst, dim, reverse)</ p > </ td >
505+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_bitand " title ="tilelang.language.reduce.reduce_bitand "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> reduce_bitand</ span > </ code > </ a > (buffer, out[, dim, clear])</ p > </ td >
506+ < td > < p > Perform reduce bitwise-and on input buffer, store the result to output buffer.</ p > </ td >
507+ </ tr >
508+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_bitor " title ="tilelang.language.reduce.reduce_bitor "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> reduce_bitor</ span > </ code > </ a > (buffer, out[, dim, clear])</ p > </ td >
509+ < td > < p > Perform reduce bitwise-or on input buffer, store the result to output buffer.</ p > </ td >
510+ </ tr >
511+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_bitxor " title ="tilelang.language.reduce.reduce_bitxor "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> reduce_bitxor</ span > </ code > </ a > (buffer, out[, dim, clear])</ p > </ td >
512+ < td > < p > Perform reduce bitwise-xor on input buffer, store the result to output buffer.</ p > </ td >
513+ </ tr >
514+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.cumsum_fragment " title ="tilelang.language.reduce.cumsum_fragment "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> cumsum_fragment</ span > </ code > </ a > (src, dst, dim, reverse)</ p > </ td >
506515< td > < p > </ p > </ td >
507516</ tr >
508- < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.cumsum " title ="tilelang.language.reduce.cumsum "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> cumsum</ span > </ code > </ a > (src[, dst, dim, reverse])</ p > </ td >
517+ < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.cumsum " title ="tilelang.language.reduce.cumsum "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> cumsum</ span > </ code > </ a > (src[, dst, dim, reverse])</ p > </ td >
509518< td > < p > Compute the cumulative sum of < cite > src</ cite > along < cite > dim</ cite > , writing results to < cite > dst</ cite > .</ p > </ td >
510519</ tr >
511- < tr class ="row-odd "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.finalize_reducer " title ="tilelang.language.reduce.finalize_reducer "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> finalize_reducer</ span > </ code > </ a > (reducer)</ p > </ td >
520+ < tr class ="row-even "> < td > < p > < a class ="reference internal " href ="#tilelang.language.reduce.finalize_reducer " title ="tilelang.language.reduce.finalize_reducer "> < code class ="xref py py-obj docutils literal notranslate "> < span class ="pre "> finalize_reducer</ span > </ code > </ a > (reducer)</ p > </ td >
512521< td > < p > Finalize a reducer buffer by emitting the < cite > tl.finalize_reducer</ cite > intrinsic.</ p > </ td >
513522</ tr >
514523</ tbody >
@@ -674,6 +683,72 @@ <h2>Module Contents<a class="headerlink" href="#module-contents" title="Link to
674683</ dl >
675684</ dd > </ dl >
676685
686+ < dl class ="py function ">
687+ < dt class ="sig sig-object py " id ="tilelang.language.reduce.reduce_bitand ">
688+ < span class ="sig-prename descclassname "> < span class ="pre "> tilelang.language.reduce.</ span > </ span > < span class ="sig-name descname "> < span class ="pre "> reduce_bitand</ span > </ span > < span class ="sig-paren "> (</ span > < em class ="sig-param "> < span class ="n "> < span class ="pre "> buffer</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> out</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> dim</ span > </ span > < span class ="o "> < span class ="pre "> =</ span > </ span > < span class ="default_value "> < span class ="pre "> -1</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> clear</ span > </ span > < span class ="o "> < span class ="pre "> =</ span > </ span > < span class ="default_value "> < span class ="pre "> True</ span > </ span > </ em > < span class ="sig-paren "> )</ span > < a class ="headerlink " href ="#tilelang.language.reduce.reduce_bitand " title ="Link to this definition "> ¶</ a > </ dt >
689+ < dd > < p > Perform reduce bitwise-and on input buffer, store the result to output buffer.</ p >
690+ < dl class ="field-list simple ">
691+ < dt class ="field-odd "> Parameters< span class ="colon "> :</ span > </ dt >
692+ < dd class ="field-odd "> < ul class ="simple ">
693+ < li > < p > < strong > buffer</ strong > (< em > tir.Buffer</ em > ) – The input buffer</ p > </ li >
694+ < li > < p > < strong > out</ strong > (< em > tir.Buffer</ em > ) – The output buffer</ p > </ li >
695+ < li > < p > < strong > dim</ strong > (< em > int</ em > ) – The dimension to perform reduce on</ p > </ li >
696+ < li > < p > < strong > clear</ strong > (< em > bool</ em > )</ p > </ li >
697+ </ ul >
698+ </ dd >
699+ < dt class ="field-even "> Returns< span class ="colon "> :</ span > </ dt >
700+ < dd class ="field-even "> < p > Handle to the reduction operation</ p >
701+ </ dd >
702+ < dt class ="field-odd "> Return type< span class ="colon "> :</ span > </ dt >
703+ < dd class ="field-odd "> < p > tir.Call</ p >
704+ </ dd >
705+ </ dl >
706+ </ dd > </ dl >
707+
708+ < dl class ="py function ">
709+ < dt class ="sig sig-object py " id ="tilelang.language.reduce.reduce_bitor ">
710+ < span class ="sig-prename descclassname "> < span class ="pre "> tilelang.language.reduce.</ span > </ span > < span class ="sig-name descname "> < span class ="pre "> reduce_bitor</ span > </ span > < span class ="sig-paren "> (</ span > < em class ="sig-param "> < span class ="n "> < span class ="pre "> buffer</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> out</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> dim</ span > </ span > < span class ="o "> < span class ="pre "> =</ span > </ span > < span class ="default_value "> < span class ="pre "> -1</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> clear</ span > </ span > < span class ="o "> < span class ="pre "> =</ span > </ span > < span class ="default_value "> < span class ="pre "> True</ span > </ span > </ em > < span class ="sig-paren "> )</ span > < a class ="headerlink " href ="#tilelang.language.reduce.reduce_bitor " title ="Link to this definition "> ¶</ a > </ dt >
711+ < dd > < p > Perform reduce bitwise-or on input buffer, store the result to output buffer.</ p >
712+ < dl class ="field-list simple ">
713+ < dt class ="field-odd "> Parameters< span class ="colon "> :</ span > </ dt >
714+ < dd class ="field-odd "> < ul class ="simple ">
715+ < li > < p > < strong > buffer</ strong > (< em > tir.Buffer</ em > ) – The input buffer</ p > </ li >
716+ < li > < p > < strong > out</ strong > (< em > tir.Buffer</ em > ) – The output buffer</ p > </ li >
717+ < li > < p > < strong > dim</ strong > (< em > int</ em > ) – The dimension to perform reduce on</ p > </ li >
718+ < li > < p > < strong > clear</ strong > (< em > bool</ em > )</ p > </ li >
719+ </ ul >
720+ </ dd >
721+ < dt class ="field-even "> Returns< span class ="colon "> :</ span > </ dt >
722+ < dd class ="field-even "> < p > Handle to the reduction operation</ p >
723+ </ dd >
724+ < dt class ="field-odd "> Return type< span class ="colon "> :</ span > </ dt >
725+ < dd class ="field-odd "> < p > tir.Call</ p >
726+ </ dd >
727+ </ dl >
728+ </ dd > </ dl >
729+
730+ < dl class ="py function ">
731+ < dt class ="sig sig-object py " id ="tilelang.language.reduce.reduce_bitxor ">
732+ < span class ="sig-prename descclassname "> < span class ="pre "> tilelang.language.reduce.</ span > </ span > < span class ="sig-name descname "> < span class ="pre "> reduce_bitxor</ span > </ span > < span class ="sig-paren "> (</ span > < em class ="sig-param "> < span class ="n "> < span class ="pre "> buffer</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> out</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> dim</ span > </ span > < span class ="o "> < span class ="pre "> =</ span > </ span > < span class ="default_value "> < span class ="pre "> -1</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> clear</ span > </ span > < span class ="o "> < span class ="pre "> =</ span > </ span > < span class ="default_value "> < span class ="pre "> True</ span > </ span > </ em > < span class ="sig-paren "> )</ span > < a class ="headerlink " href ="#tilelang.language.reduce.reduce_bitxor " title ="Link to this definition "> ¶</ a > </ dt >
733+ < dd > < p > Perform reduce bitwise-xor on input buffer, store the result to output buffer.</ p >
734+ < dl class ="field-list simple ">
735+ < dt class ="field-odd "> Parameters< span class ="colon "> :</ span > </ dt >
736+ < dd class ="field-odd "> < ul class ="simple ">
737+ < li > < p > < strong > buffer</ strong > (< em > tir.Buffer</ em > ) – The input buffer</ p > </ li >
738+ < li > < p > < strong > out</ strong > (< em > tir.Buffer</ em > ) – The output buffer</ p > </ li >
739+ < li > < p > < strong > dim</ strong > (< em > int</ em > ) – The dimension to perform reduce on</ p > </ li >
740+ < li > < p > < strong > clear</ strong > (< em > bool</ em > )</ p > </ li >
741+ </ ul >
742+ </ dd >
743+ < dt class ="field-even "> Returns< span class ="colon "> :</ span > </ dt >
744+ < dd class ="field-even "> < p > Handle to the reduction operation</ p >
745+ </ dd >
746+ < dt class ="field-odd "> Return type< span class ="colon "> :</ span > </ dt >
747+ < dd class ="field-odd "> < p > tir.Call</ p >
748+ </ dd >
749+ </ dl >
750+ </ dd > </ dl >
751+
677752< dl class ="py function ">
678753< dt class ="sig sig-object py " id ="tilelang.language.reduce.cumsum_fragment ">
679754< span class ="sig-prename descclassname "> < span class ="pre "> tilelang.language.reduce.</ span > </ span > < span class ="sig-name descname "> < span class ="pre "> cumsum_fragment</ span > </ span > < span class ="sig-paren "> (</ span > < em class ="sig-param "> < span class ="n "> < span class ="pre "> src</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> dst</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> dim</ span > </ span > </ em > , < em class ="sig-param "> < span class ="n "> < span class ="pre "> reverse</ span > </ span > </ em > < span class ="sig-paren "> )</ span > < a class ="headerlink " href ="#tilelang.language.reduce.cumsum_fragment " title ="Link to this definition "> ¶</ a > </ dt >
@@ -824,6 +899,9 @@ <h2>Module Contents<a class="headerlink" href="#module-contents" title="Link to
824899< li > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_sum "> < code class ="docutils literal notranslate "> < span class ="pre "> reduce_sum()</ span > </ code > </ a > </ li >
825900< li > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_abssum "> < code class ="docutils literal notranslate "> < span class ="pre "> reduce_abssum()</ span > </ code > </ a > </ li >
826901< li > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_absmax "> < code class ="docutils literal notranslate "> < span class ="pre "> reduce_absmax()</ span > </ code > </ a > </ li >
902+ < li > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_bitand "> < code class ="docutils literal notranslate "> < span class ="pre "> reduce_bitand()</ span > </ code > </ a > </ li >
903+ < li > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_bitor "> < code class ="docutils literal notranslate "> < span class ="pre "> reduce_bitor()</ span > </ code > </ a > </ li >
904+ < li > < a class ="reference internal " href ="#tilelang.language.reduce.reduce_bitxor "> < code class ="docutils literal notranslate "> < span class ="pre "> reduce_bitxor()</ span > </ code > </ a > </ li >
827905< li > < a class ="reference internal " href ="#tilelang.language.reduce.cumsum_fragment "> < code class ="docutils literal notranslate "> < span class ="pre "> cumsum_fragment()</ span > </ code > </ a > </ li >
828906< li > < a class ="reference internal " href ="#tilelang.language.reduce.cumsum "> < code class ="docutils literal notranslate "> < span class ="pre "> cumsum()</ span > </ code > </ a > </ li >
829907< li > < a class ="reference internal " href ="#tilelang.language.reduce.finalize_reducer "> < code class ="docutils literal notranslate "> < span class ="pre "> finalize_reducer()</ span > </ code > </ a > </ li >
0 commit comments