remove mkl::remainder implementation

Author: Vahid Tavanashad

dpnp/backend/extensions/vm/remainder.hpp

@@ -68,21 +68,6 @@ struct DivOutputType
         dpctl_td_ns::DefaultResultEntry<void>>::result_type;
 };
 
-/**
- * @brief A factory to define pairs of supported types for which
- * MKL VM library provides support in oneapi::mkl::vm::remainder<T> function.
- *
- * @tparam T Type of input vectors `a` and `b` and of result vector `y`.
- */
-template <typename T>
-struct RemainderOutputType
-{
-    using value_type = typename std::disjunction<
-        dpctl_td_ns::BinaryTypeMapResultEntry<T, double, T, double, double>,
-        dpctl_td_ns::BinaryTypeMapResultEntry<T, float, T, float, float>,
-        dpctl_td_ns::DefaultResultEntry<void>>::result_type;
-};
-
 /**
  * @brief A factory to define pairs of supported types for which
  * MKL VM library provides support in oneapi::mkl::vm::cos<T> function.

dpnp/backend/extensions/vm/types_matrix.hpp

@@ -34,7 +34,6 @@
 #include "cos.hpp"
 #include "div.hpp"
 #include "ln.hpp"
-#include "remainder.hpp"
 #include "sin.hpp"
 #include "sqr.hpp"
 #include "sqrt.hpp"
@@ -47,7 +46,6 @@ using vm_ext::binary_impl_fn_ptr_t;
 using vm_ext::unary_impl_fn_ptr_t;
 
 static binary_impl_fn_ptr_t div_dispatch_vector[dpctl_td_ns::num_types];
-static binary_impl_fn_ptr_t remainder_dispatch_vector[dpctl_td_ns::num_types];
 
 static unary_impl_fn_ptr_t cos_dispatch_vector[dpctl_td_ns::num_types];
 static unary_impl_fn_ptr_t ln_dispatch_vector[dpctl_td_ns::num_types];
@@ -90,37 +88,6 @@ PYBIND11_MODULE(_vm_impl, m)
               py::arg("dst"));
     }
 
-    // BinaryUfunc: ==== REMAINDER(x1, x2) ====
-    {
-        vm_ext::init_ufunc_dispatch_vector<binary_impl_fn_ptr_t,
-                                           vm_ext::RemainderContigFactory>(
-            remainder_dispatch_vector);
-
-        auto remainder_pyapi = [&](sycl::queue exec_q, arrayT src1, arrayT src2,
-                                   arrayT dst, const event_vecT &depends = {}) {
-            return vm_ext::binary_ufunc(exec_q, src1, src2, dst, depends,
-                                        remainder_dispatch_vector);
-        };
-        m.def("_remainder", remainder_pyapi,
-              "Call `remainder` function from OneMKL VM library to performs "
-              "element "
-              "by element remainder of vector `src1` by vector `src2` "
-              "to resulting vector `dst`",
-              py::arg("sycl_queue"), py::arg("src1"), py::arg("src2"),
-              py::arg("dst"), py::arg("depends") = py::list());
-
-        auto remainder_need_to_call_pyapi = [&](sycl::queue exec_q, arrayT src1,
-                                                arrayT src2, arrayT dst) {
-            return vm_ext::need_to_call_binary_ufunc(exec_q, src1, src2, dst,
-                                                     remainder_dispatch_vector);
-        };
-        m.def("_mkl_remainder_to_call", remainder_need_to_call_pyapi,
-              "Check input arguments to answer if `remainder` function from "
-              "OneMKL VM library can be used",
-              py::arg("sycl_queue"), py::arg("src1"), py::arg("src2"),
-              py::arg("dst"));
-    }
-
     // UnaryUfunc: ==== Cos(x) ====
     {
         vm_ext::init_ufunc_dispatch_vector<unary_impl_fn_ptr_t,

dpnp/backend/extensions/vm/vm_py.cpp

@@ -980,24 +980,6 @@ def dpnp_not_equal(x1, x2, out=None, order="K"):
 
 
 def dpnp_remainder(x1, x2, out=None, order="K"):
-    """
-    Invokes remainder() function from pybind11 extension of OneMKL VM if possible.
-
-    Otherwise fully relies on dpctl.tensor implementation for remainder() function.
-
-    """
-
-    def _call_remainder(src1, src2, dst, sycl_queue, depends=None):
-        """A callback to register in BinaryElementwiseFunc class of dpctl.tensor"""
-
-        if depends is None:
-            depends = []
-
-        if vmi._mkl_remainder_to_call(sycl_queue, src1, src2, dst):
-            # call pybind11 extension for remainder() function from OneMKL VM
-            return vmi._remainder(sycl_queue, src1, src2, dst, depends)
-        return ti._remainder(src1, src2, dst, sycl_queue, depends)
-
     # dpctl.tensor only works with usm_ndarray or scalar
     x1_usm_or_scalar = dpnp.get_usm_ndarray_or_scalar(x1)
     x2_usm_or_scalar = dpnp.get_usm_ndarray_or_scalar(x2)
@@ -1006,7 +988,7 @@ def _call_remainder(src1, src2, dst, sycl_queue, depends=None):
     func = BinaryElementwiseFunc(
         "remainder",
         ti._remainder_result_type,
-        _call_remainder,
+        ti._remainder,
         _remainder_docstring_,
     )
     res_usm = func(x1_usm_or_scalar, x2_usm_or_scalar, out=out_usm, order=order)

dpnp/dpnp_algo/dpnp_elementwise_common.py

@@ -1602,12 +1602,12 @@ def remainder(
 
     Limitations
     -----------
-        Parameters ``x1`` and ``x2`` are supported as either :obj:`dpnp.ndarray` or scalar.
-        Parameters ``dtype``, ``out`` and ``where`` are supported with their default values.
-        Keyword arguments ``kwargs`` are currently unsupported.
+        Parameters ``x1`` and ``x2`` are supported as either :obj:`dpnp.ndarray`,
+        :class:`dpctl.tensor.usm_ndarray` or scalar.
+        Parameters `where`, `dtype` and `subok` are supported with their default values.
         Otherwise the functions will be executed sequentially on CPU.
+        Keyword arguments `kwargs` are currently unsupported.
         Input array data types are limited by supported DPNP :ref:`Data types`.
-        Parameters ``x1`` and ``x2`` are supported with equal sizes and shapes.
 
     See Also
     --------