Update blas extension to use init_dispatch_vector() and using init_dispatch_table() from common utils

Author: antonwolfy

dpnp/backend/extensions/blas/CMakeLists.txt

@@ -60,8 +60,7 @@ endif()
 
 set_target_properties(${python_module_name} PROPERTIES CMAKE_POSITION_INDEPENDENT_CODE ON)
 
-target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../../include)
-target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../../src)
+target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../../)
 target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../common)
 
 target_include_directories(${python_module_name} PUBLIC ${Dpctl_INCLUDE_DIRS})

dpnp/backend/extensions/blas/blas_py.cpp

@@ -30,6 +30,9 @@
 #include <pybind11/pybind11.h>
 #include <pybind11/stl.h>
 
+// utils extension header
+#include "ext/common.hpp"
+
 #include "dot.hpp"
 #include "dot_common.hpp"
 #include "dotc.hpp"
@@ -41,7 +44,9 @@
 namespace blas_ns = dpnp::extensions::blas;
 namespace py = pybind11;
 namespace dot_ns = blas_ns::dot;
+
 using dot_ns::dot_impl_fn_ptr_t;
+using ext::common::init_dispatch_vector;
 
 // populate dispatch vectors and tables
 void init_dispatch_vectors_tables(void)
@@ -64,7 +69,7 @@ PYBIND11_MODULE(_blas_impl, m)
     using event_vecT = std::vector<sycl::event>;
 
     {
-        dot_ns::init_dot_dispatch_vector<blas_ns::DotContigFactory>(
+        init_dispatch_vector<dot_impl_fn_ptr_t, blas_ns::DotContigFactory>(
             dot_dispatch_vector);
 
         auto dot_pyapi = [&](sycl::queue &exec_q, const arrayT &src1,
@@ -82,7 +87,7 @@ PYBIND11_MODULE(_blas_impl, m)
     }
 
     {
-        dot_ns::init_dot_dispatch_vector<blas_ns::DotcContigFactory>(
+        init_dispatch_vector<dot_impl_fn_ptr_t, blas_ns::DotcContigFactory>(
             dotc_dispatch_vector);
 
         auto dotc_pyapi = [&](sycl::queue &exec_q, const arrayT &src1,
@@ -101,7 +106,7 @@ PYBIND11_MODULE(_blas_impl, m)
     }
 
     {
-        dot_ns::init_dot_dispatch_vector<blas_ns::DotuContigFactory>(
+        init_dispatch_vector<dot_impl_fn_ptr_t, blas_ns::DotuContigFactory>(
             dotu_dispatch_vector);
 
         auto dotu_pyapi = [&](sycl::queue &exec_q, const arrayT &src1,

dpnp/backend/extensions/blas/dot_common.hpp

@@ -165,13 +165,4 @@ std::pair<sycl::event, sycl::event>
 
     return std::make_pair(args_ev, dot_ev);
 }
-
-template <template <typename fnT, typename T> typename factoryT>
-void init_dot_dispatch_vector(dot_impl_fn_ptr_t dot_dispatch_vector[])
-{
-    dpctl_td_ns::DispatchVectorBuilder<dot_impl_fn_ptr_t, factoryT,
-                                       dpctl_td_ns::num_types>
-        contig;
-    contig.populate_dispatch_vector(dot_dispatch_vector);
-}
 } // namespace dpnp::extensions::blas::dot

dpnp/backend/extensions/blas/gemm.cpp

@@ -27,6 +27,9 @@
 
 #include <pybind11/pybind11.h>
 
+// utils extension header
+#include "ext/common.hpp"
+
 // dpctl tensor headers
 #include "utils/memory_overlap.hpp"
 #include "utils/output_validation.hpp"
@@ -35,14 +38,14 @@
 #include "gemm.hpp"
 #include "types_matrix.hpp"
 
-#include "dpnp_utils.hpp"
-
 namespace dpnp::extensions::blas
 {
 namespace mkl_blas = oneapi::mkl::blas;
 namespace py = pybind11;
 namespace type_utils = dpctl::tensor::type_utils;
 
+using ext::common::init_dispatch_table;
+
 typedef sycl::event (*gemm_impl_fn_ptr_t)(sycl::queue &,
                                           oneapi::mkl::transpose,
                                           oneapi::mkl::transpose,
@@ -339,9 +342,7 @@ struct GemmContigFactory
 
 void init_gemm_dispatch_table(void)
 {
-    dpctl_td_ns::DispatchTableBuilder<gemm_impl_fn_ptr_t, GemmContigFactory,
-                                      dpctl_td_ns::num_types>
-        contig;
-    contig.populate_dispatch_table(gemm_dispatch_table);
+    init_dispatch_table<gemm_impl_fn_ptr_t, GemmContigFactory>(
+        gemm_dispatch_table);
 }
 } // namespace dpnp::extensions::blas

dpnp/backend/extensions/blas/gemm_batch.cpp

@@ -27,6 +27,9 @@
 
 #include <pybind11/pybind11.h>
 
+// utils extension header
+#include "ext/common.hpp"
+
 // dpctl tensor headers
 #include "utils/memory_overlap.hpp"
 #include "utils/output_validation.hpp"
@@ -35,14 +38,14 @@
 #include "gemm.hpp"
 #include "types_matrix.hpp"
 
-#include "dpnp_utils.hpp"
-
 namespace dpnp::extensions::blas
 {
 namespace mkl_blas = oneapi::mkl::blas;
 namespace py = pybind11;
 namespace type_utils = dpctl::tensor::type_utils;
 
+using ext::common::init_dispatch_table;
+
 typedef sycl::event (*gemm_batch_impl_fn_ptr_t)(
     sycl::queue &,
     const std::int64_t,
@@ -415,10 +418,7 @@ struct GemmBatchContigFactory
 
 void init_gemm_batch_dispatch_table(void)
 {
-    dpctl_td_ns::DispatchTableBuilder<gemm_batch_impl_fn_ptr_t,
-                                      GemmBatchContigFactory,
-                                      dpctl_td_ns::num_types>
-        contig;
-    contig.populate_dispatch_table(gemm_batch_dispatch_table);
+    init_dispatch_table<gemm_batch_impl_fn_ptr_t, GemmBatchContigFactory>(
+        gemm_batch_dispatch_table);
 }
 } // namespace dpnp::extensions::blas

dpnp/backend/extensions/blas/gemv.cpp

@@ -27,6 +27,9 @@
 
 #include <pybind11/pybind11.h>
 
+// utils extension header
+#include "ext/common.hpp"
+
 // dpctl tensor headers
 #include "utils/memory_overlap.hpp"
 #include "utils/output_validation.hpp"
@@ -35,14 +38,14 @@
 #include "gemv.hpp"
 #include "types_matrix.hpp"
 
-#include "dpnp_utils.hpp"
-
 namespace dpnp::extensions::blas
 {
 namespace mkl_blas = oneapi::mkl::blas;
 namespace py = pybind11;
 namespace type_utils = dpctl::tensor::type_utils;
 
+using ext::common::init_dispatch_vector;
+
 typedef sycl::event (*gemv_impl_fn_ptr_t)(sycl::queue &,
                                           oneapi::mkl::transpose,
                                           const std::int64_t,
@@ -320,9 +323,7 @@ struct GemvContigFactory
 
 void init_gemv_dispatch_vector(void)
 {
-    dpctl_td_ns::DispatchVectorBuilder<gemv_impl_fn_ptr_t, GemvContigFactory,
-                                       dpctl_td_ns::num_types>
-        contig;
-    contig.populate_dispatch_vector(gemv_dispatch_vector);
+    init_dispatch_vector<gemv_impl_fn_ptr_t, GemvContigFactory>(
+        gemv_dispatch_vector);
 }
 } // namespace dpnp::extensions::blas

dpnp/backend/extensions/blas/syrk.cpp

@@ -28,6 +28,7 @@
 
 #include <pybind11/pybind11.h>
 
+// utils extension header
 #include "ext/common.hpp"
 
 // dpctl tensor headers
@@ -38,8 +39,6 @@
 #include "syrk.hpp"
 #include "types_matrix.hpp"
 
-#include "dpnp_utils.hpp"
-
 using ext::common::Align;
 
 namespace dpnp::extensions::blas
@@ -48,6 +47,8 @@ namespace mkl_blas = oneapi::mkl::blas;
 namespace py = pybind11;
 namespace type_utils = dpctl::tensor::type_utils;
 
+using ext::common::init_dispatch_vector;
+
 typedef sycl::event (*syrk_impl_fn_ptr_t)(sycl::queue &,
                                           const oneapi::mkl::transpose,
                                           const std::int64_t,
@@ -349,9 +350,7 @@ struct SyrkContigFactory
 
 void init_syrk_dispatch_vector(void)
 {
-    dpctl_td_ns::DispatchVectorBuilder<syrk_impl_fn_ptr_t, SyrkContigFactory,
-                                       dpctl_td_ns::num_types>
-        contig;
-    contig.populate_dispatch_vector(syrk_dispatch_vector);
+    init_dispatch_vector<syrk_impl_fn_ptr_t, SyrkContigFactory>(
+        syrk_dispatch_vector);
 }
 } // namespace dpnp::extensions::blas