From 77840e6e13e91efb3c8266e7686c39708c563125 Mon Sep 17 00:00:00 2001 From: Vahid Tavanashad Date: Mon, 14 Aug 2023 13:15:30 -0500 Subject: [PATCH 1/4] use_dpctl_ceil_floor_trunc_in_dpnp --- dpnp/backend/extensions/vm/ceil.hpp | 78 ++++ dpnp/backend/extensions/vm/floor.hpp | 78 ++++ dpnp/backend/extensions/vm/trunc.hpp | 78 ++++ dpnp/backend/extensions/vm/types_matrix.hpp | 46 ++ dpnp/backend/extensions/vm/vm_py.cpp | 90 ++++ dpnp/backend/include/dpnp_iface_fptr.hpp | 8 +- dpnp/backend/kernels/dpnp_krnl_elemwise.cpp | 27 -- dpnp/dpnp_algo/dpnp_algo.pxd | 6 - dpnp/dpnp_algo/dpnp_algo_mathematical.pxi | 15 - dpnp/dpnp_algo/dpnp_elementwise_common.py | 164 ++++++++ dpnp/dpnp_iface_mathematical.py | 392 ++++++++++-------- dpnp/dpnp_iface_trigonometric.py | 6 +- tests/skipped_tests.tbl | 3 - tests/skipped_tests_gpu.tbl | 8 - .../cupy/math_tests/test_rounding.py | 2 +- 15 files changed, 765 insertions(+), 236 deletions(-) create mode 100644 dpnp/backend/extensions/vm/ceil.hpp create mode 100644 dpnp/backend/extensions/vm/floor.hpp create mode 100644 dpnp/backend/extensions/vm/trunc.hpp diff --git a/dpnp/backend/extensions/vm/ceil.hpp b/dpnp/backend/extensions/vm/ceil.hpp new file mode 100644 index 00000000000..b1354b884c3 --- /dev/null +++ b/dpnp/backend/extensions/vm/ceil.hpp @@ -0,0 +1,78 @@ +//***************************************************************************** +// Copyright (c) 2023, Intel Corporation +// All rights reserved. +// +// Redistribution and use in source and binary forms, with or without +// modification, are permitted provided that the following conditions are met: +// - Redistributions of source code must retain the above copyright notice, +// this list of conditions and the following disclaimer. +// - Redistributions in binary form must reproduce the above copyright notice, +// this list of conditions and the following disclaimer in the documentation +// and/or other materials provided with the distribution. +// +// THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" +// AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE +// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE +// ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE +// LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR +// CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF +// SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS +// INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN +// CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) +// ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF +// THE POSSIBILITY OF SUCH DAMAGE. +//***************************************************************************** + +#pragma once + +#include + +#include "common.hpp" +#include "types_matrix.hpp" + +namespace dpnp +{ +namespace backend +{ +namespace ext +{ +namespace vm +{ +template +sycl::event ceil_contig_impl(sycl::queue exec_q, + const std::int64_t n, + const char *in_a, + char *out_y, + const std::vector &depends) +{ + type_utils::validate_type_for_device(exec_q); + + const T *a = reinterpret_cast(in_a); + T *y = reinterpret_cast(out_y); + + return mkl_vm::ceil(exec_q, + n, // number of elements to be calculated + a, // pointer `a` containing input vector of size n + y, // pointer `y` to the output vector of size n + depends); +} + +template +struct CeilContigFactory +{ + fnT get() + { + if constexpr (std::is_same_v< + typename types::CeilOutputType::value_type, void>) + { + return nullptr; + } + else { + return ceil_contig_impl; + } + } +}; +} // namespace vm +} // namespace ext +} // namespace backend +} // namespace dpnp diff --git a/dpnp/backend/extensions/vm/floor.hpp b/dpnp/backend/extensions/vm/floor.hpp new file mode 100644 index 00000000000..66707ba2979 --- /dev/null +++ b/dpnp/backend/extensions/vm/floor.hpp @@ -0,0 +1,78 @@ +//***************************************************************************** +// Copyright (c) 2023, Intel Corporation +// All rights reserved. +// +// Redistribution and use in source and binary forms, with or without +// modification, are permitted provided that the following conditions are met: +// - Redistributions of source code must retain the above copyright notice, +// this list of conditions and the following disclaimer. +// - Redistributions in binary form must reproduce the above copyright notice, +// this list of conditions and the following disclaimer in the documentation +// and/or other materials provided with the distribution. +// +// THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" +// AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE +// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE +// ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE +// LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR +// CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF +// SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS +// INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN +// CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) +// ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF +// THE POSSIBILITY OF SUCH DAMAGE. +//***************************************************************************** + +#pragma once + +#include + +#include "common.hpp" +#include "types_matrix.hpp" + +namespace dpnp +{ +namespace backend +{ +namespace ext +{ +namespace vm +{ +template +sycl::event floor_contig_impl(sycl::queue exec_q, + const std::int64_t n, + const char *in_a, + char *out_y, + const std::vector &depends) +{ + type_utils::validate_type_for_device(exec_q); + + const T *a = reinterpret_cast(in_a); + T *y = reinterpret_cast(out_y); + + return mkl_vm::floor(exec_q, + n, // number of elements to be calculated + a, // pointer `a` containing input vector of size n + y, // pointer `y` to the output vector of size n + depends); +} + +template +struct FloorContigFactory +{ + fnT get() + { + if constexpr (std::is_same_v< + typename types::FloorOutputType::value_type, void>) + { + return nullptr; + } + else { + return floor_contig_impl; + } + } +}; +} // namespace vm +} // namespace ext +} // namespace backend +} // namespace dpnp diff --git a/dpnp/backend/extensions/vm/trunc.hpp b/dpnp/backend/extensions/vm/trunc.hpp new file mode 100644 index 00000000000..d8ecadff363 --- /dev/null +++ b/dpnp/backend/extensions/vm/trunc.hpp @@ -0,0 +1,78 @@ +//***************************************************************************** +// Copyright (c) 2023, Intel Corporation +// All rights reserved. +// +// Redistribution and use in source and binary forms, with or without +// modification, are permitted provided that the following conditions are met: +// - Redistributions of source code must retain the above copyright notice, +// this list of conditions and the following disclaimer. +// - Redistributions in binary form must reproduce the above copyright notice, +// this list of conditions and the following disclaimer in the documentation +// and/or other materials provided with the distribution. +// +// THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" +// AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE +// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE +// ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE +// LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR +// CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF +// SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS +// INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN +// CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) +// ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF +// THE POSSIBILITY OF SUCH DAMAGE. +//***************************************************************************** + +#pragma once + +#include + +#include "common.hpp" +#include "types_matrix.hpp" + +namespace dpnp +{ +namespace backend +{ +namespace ext +{ +namespace vm +{ +template +sycl::event trunc_contig_impl(sycl::queue exec_q, + const std::int64_t n, + const char *in_a, + char *out_y, + const std::vector &depends) +{ + type_utils::validate_type_for_device(exec_q); + + const T *a = reinterpret_cast(in_a); + T *y = reinterpret_cast(out_y); + + return mkl_vm::trunc(exec_q, + n, // number of elements to be calculated + a, // pointer `a` containing input vector of size n + y, // pointer `y` to the output vector of size n + depends); +} + +template +struct TruncContigFactory +{ + fnT get() + { + if constexpr (std::is_same_v< + typename types::TruncOutputType::value_type, void>) + { + return nullptr; + } + else { + return trunc_contig_impl; + } + } +}; +} // namespace vm +} // namespace ext +} // namespace backend +} // namespace dpnp diff --git a/dpnp/backend/extensions/vm/types_matrix.hpp b/dpnp/backend/extensions/vm/types_matrix.hpp index 667e27f0c9a..cd4fd76d4be 100644 --- a/dpnp/backend/extensions/vm/types_matrix.hpp +++ b/dpnp/backend/extensions/vm/types_matrix.hpp @@ -68,6 +68,21 @@ struct DivOutputType dpctl_td_ns::DefaultResultEntry>::result_type; }; +/** + * @brief A factory to define pairs of supported types for which + * MKL VM library provides support in oneapi::mkl::vm::ceil function. + * + * @tparam T Type of input vector `a` and of result vector `y`. + */ +template +struct CeilOutputType +{ + using value_type = typename std::disjunction< + dpctl_td_ns::TypeMapResultEntry, + dpctl_td_ns::TypeMapResultEntry, + dpctl_td_ns::DefaultResultEntry>::result_type; +}; + /** * @brief A factory to define pairs of supported types for which * MKL VM library provides support in oneapi::mkl::vm::cos function. @@ -87,6 +102,21 @@ struct CosOutputType dpctl_td_ns::DefaultResultEntry>::result_type; }; +/** + * @brief A factory to define pairs of supported types for which + * MKL VM library provides support in oneapi::mkl::vm::floor function. + * + * @tparam T Type of input vector `a` and of result vector `y`. + */ +template +struct FloorOutputType +{ + using value_type = typename std::disjunction< + dpctl_td_ns::TypeMapResultEntry, + dpctl_td_ns::TypeMapResultEntry, + dpctl_td_ns::DefaultResultEntry>::result_type; +}; + /** * @brief A factory to define pairs of supported types for which * MKL VM library provides support in oneapi::mkl::vm::ln function. @@ -158,6 +188,22 @@ struct SqrtOutputType dpctl_td_ns::TypeMapResultEntry, dpctl_td_ns::DefaultResultEntry>::result_type; }; + +/** + * @brief A factory to define pairs of supported types for which + * MKL VM library provides support in oneapi::mkl::vm::trunc function. + * + * @tparam T Type of input vector `a` and of result vector `y`. + */ +template +struct TruncOutputType +{ + using value_type = typename std::disjunction< + dpctl_td_ns::TypeMapResultEntry, + dpctl_td_ns::TypeMapResultEntry, + dpctl_td_ns::DefaultResultEntry>::result_type; +}; + } // namespace types } // namespace vm } // namespace ext diff --git a/dpnp/backend/extensions/vm/vm_py.cpp b/dpnp/backend/extensions/vm/vm_py.cpp index 924db66025b..c7435ae9e2e 100644 --- a/dpnp/backend/extensions/vm/vm_py.cpp +++ b/dpnp/backend/extensions/vm/vm_py.cpp @@ -30,13 +30,16 @@ #include #include +#include "ceil.hpp" #include "common.hpp" #include "cos.hpp" #include "div.hpp" +#include "floor.hpp" #include "ln.hpp" #include "sin.hpp" #include "sqr.hpp" #include "sqrt.hpp" +#include "trunc.hpp" #include "types_matrix.hpp" namespace py = pybind11; @@ -47,11 +50,14 @@ using vm_ext::unary_impl_fn_ptr_t; static binary_impl_fn_ptr_t div_dispatch_vector[dpctl_td_ns::num_types]; +static unary_impl_fn_ptr_t ceil_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 floor_dispatch_vector[dpctl_td_ns::num_types]; static unary_impl_fn_ptr_t ln_dispatch_vector[dpctl_td_ns::num_types]; static unary_impl_fn_ptr_t sin_dispatch_vector[dpctl_td_ns::num_types]; static unary_impl_fn_ptr_t sqr_dispatch_vector[dpctl_td_ns::num_types]; static unary_impl_fn_ptr_t sqrt_dispatch_vector[dpctl_td_ns::num_types]; +static unary_impl_fn_ptr_t trunc_dispatch_vector[dpctl_td_ns::num_types]; PYBIND11_MODULE(_vm_impl, m) { @@ -88,6 +94,34 @@ PYBIND11_MODULE(_vm_impl, m) py::arg("dst")); } + // UnaryUfunc: ==== Ceil(x) ==== + { + vm_ext::init_ufunc_dispatch_vector( + ceil_dispatch_vector); + + auto ceil_pyapi = [&](sycl::queue exec_q, arrayT src, arrayT dst, + const event_vecT &depends = {}) { + return vm_ext::unary_ufunc(exec_q, src, dst, depends, + ceil_dispatch_vector); + }; + m.def("_ceil", ceil_pyapi, + "Call `ceil` function from OneMKL VM library to compute " + "ceiling of vector elements", + py::arg("sycl_queue"), py::arg("src"), py::arg("dst"), + py::arg("depends") = py::list()); + + auto ceil_need_to_call_pyapi = [&](sycl::queue exec_q, arrayT src, + arrayT dst) { + return vm_ext::need_to_call_unary_ufunc(exec_q, src, dst, + ceil_dispatch_vector); + }; + m.def("_mkl_ceil_to_call", ceil_need_to_call_pyapi, + "Check input arguments to answer if `ceil` function from " + "OneMKL VM library can be used", + py::arg("sycl_queue"), py::arg("src"), py::arg("dst")); + } + // UnaryUfunc: ==== Cos(x) ==== { vm_ext::init_ufunc_dispatch_vector( + floor_dispatch_vector); + + auto floor_pyapi = [&](sycl::queue exec_q, arrayT src, arrayT dst, + const event_vecT &depends = {}) { + return vm_ext::unary_ufunc(exec_q, src, dst, depends, + floor_dispatch_vector); + }; + m.def("_floor", floor_pyapi, + "Call `floor` function from OneMKL VM library to compute " + "floor of vector elements", + py::arg("sycl_queue"), py::arg("src"), py::arg("dst"), + py::arg("depends") = py::list()); + + auto floor_need_to_call_pyapi = [&](sycl::queue exec_q, arrayT src, + arrayT dst) { + return vm_ext::need_to_call_unary_ufunc(exec_q, src, dst, + floor_dispatch_vector); + }; + m.def("_mkl_floor_to_call", floor_need_to_call_pyapi, + "Check input arguments to answer if `floor` function from " + "OneMKL VM library can be used", + py::arg("sycl_queue"), py::arg("src"), py::arg("dst")); + } + // UnaryUfunc: ==== Ln(x) ==== { vm_ext::init_ufunc_dispatch_vector( + trunc_dispatch_vector); + + auto trunc_pyapi = [&](sycl::queue exec_q, arrayT src, arrayT dst, + const event_vecT &depends = {}) { + return vm_ext::unary_ufunc(exec_q, src, dst, depends, + trunc_dispatch_vector); + }; + m.def("_trunc", trunc_pyapi, + "Call `trunc` function from OneMKL VM library to compute " + "the truncated value of vector elements", + py::arg("sycl_queue"), py::arg("src"), py::arg("dst"), + py::arg("depends") = py::list()); + + auto trunc_need_to_call_pyapi = [&](sycl::queue exec_q, arrayT src, + arrayT dst) { + return vm_ext::need_to_call_unary_ufunc(exec_q, src, dst, + trunc_dispatch_vector); + }; + m.def("_mkl_trunc_to_call", trunc_need_to_call_pyapi, + "Check input arguments to answer if `trunc` function from " + "OneMKL VM library can be used", + py::arg("sycl_queue"), py::arg("src"), py::arg("dst")); + } } diff --git a/dpnp/backend/include/dpnp_iface_fptr.hpp b/dpnp/backend/include/dpnp_iface_fptr.hpp index e2c29a0f4a8..9917eedd678 100644 --- a/dpnp/backend/include/dpnp_iface_fptr.hpp +++ b/dpnp/backend/include/dpnp_iface_fptr.hpp @@ -113,8 +113,6 @@ enum class DPNPFuncName : size_t DPNP_FN_CBRT_EXT, /**< Used in numpy.cbrt() impl, requires extra parameters */ DPNP_FN_CEIL, /**< Used in numpy.ceil() impl */ - DPNP_FN_CEIL_EXT, /**< Used in numpy.ceil() impl, requires extra parameters - */ DPNP_FN_CHOLESKY, /**< Used in numpy.linalg.cholesky() impl */ DPNP_FN_CHOLESKY_EXT, /**< Used in numpy.linalg.cholesky() impl, requires extra parameters */ @@ -212,8 +210,6 @@ enum class DPNPFuncName : size_t DPNP_FN_FLATTEN_EXT, /**< Used in numpy.flatten() impl, requires extra parameters */ DPNP_FN_FLOOR, /**< Used in numpy.floor() impl */ - DPNP_FN_FLOOR_EXT, /**< Used in numpy.floor() impl, requires extra - parameters */ DPNP_FN_FLOOR_DIVIDE, /**< Used in numpy.floor_divide() impl */ DPNP_FN_FLOOR_DIVIDE_EXT, /**< Used in numpy.floor_divide() impl, requires extra parameters */ @@ -488,9 +484,7 @@ enum class DPNPFuncName : size_t DPNP_FN_TRIL, /**< Used in numpy.tril() impl */ DPNP_FN_TRIU, /**< Used in numpy.triu() impl */ DPNP_FN_TRUNC, /**< Used in numpy.trunc() impl */ - DPNP_FN_TRUNC_EXT, /**< Used in numpy.trunc() impl, requires extra - parameters */ - DPNP_FN_VANDER, /**< Used in numpy.vander() impl */ + DPNP_FN_VANDER, /**< Used in numpy.vander() impl */ DPNP_FN_VANDER_EXT, /**< Used in numpy.vander() impl, requires extra parameters */ DPNP_FN_VAR, /**< Used in numpy.var() impl */ diff --git a/dpnp/backend/kernels/dpnp_krnl_elemwise.cpp b/dpnp/backend/kernels/dpnp_krnl_elemwise.cpp index 1bd0d1922e0..a382d85b573 100644 --- a/dpnp/backend/kernels/dpnp_krnl_elemwise.cpp +++ b/dpnp/backend/kernels/dpnp_krnl_elemwise.cpp @@ -359,15 +359,6 @@ static void func_map_init_elemwise_1arg_2type(func_map_t &fmap) fmap[DPNPFuncName::DPNP_FN_CEIL][eft_DBL][eft_DBL] = { eft_DBL, (void *)dpnp_ceil_c_default}; - fmap[DPNPFuncName::DPNP_FN_CEIL_EXT][eft_INT][eft_INT] = { - eft_DBL, (void *)dpnp_ceil_c_ext}; - fmap[DPNPFuncName::DPNP_FN_CEIL_EXT][eft_LNG][eft_LNG] = { - eft_DBL, (void *)dpnp_ceil_c_ext}; - fmap[DPNPFuncName::DPNP_FN_CEIL_EXT][eft_FLT][eft_FLT] = { - eft_FLT, (void *)dpnp_ceil_c_ext}; - fmap[DPNPFuncName::DPNP_FN_CEIL_EXT][eft_DBL][eft_DBL] = { - eft_DBL, (void *)dpnp_ceil_c_ext}; - fmap[DPNPFuncName::DPNP_FN_COPYTO][eft_BLN][eft_BLN] = { eft_BLN, (void *)dpnp_copyto_c_default}; fmap[DPNPFuncName::DPNP_FN_COPYTO][eft_BLN][eft_INT] = { @@ -603,15 +594,6 @@ static void func_map_init_elemwise_1arg_2type(func_map_t &fmap) fmap[DPNPFuncName::DPNP_FN_FLOOR][eft_DBL][eft_DBL] = { eft_DBL, (void *)dpnp_floor_c_default}; - fmap[DPNPFuncName::DPNP_FN_FLOOR_EXT][eft_INT][eft_INT] = { - eft_DBL, (void *)dpnp_floor_c_ext}; - fmap[DPNPFuncName::DPNP_FN_FLOOR_EXT][eft_LNG][eft_LNG] = { - eft_DBL, (void *)dpnp_floor_c_ext}; - fmap[DPNPFuncName::DPNP_FN_FLOOR_EXT][eft_FLT][eft_FLT] = { - eft_FLT, (void *)dpnp_floor_c_ext}; - fmap[DPNPFuncName::DPNP_FN_FLOOR_EXT][eft_DBL][eft_DBL] = { - eft_DBL, (void *)dpnp_floor_c_ext}; - fmap[DPNPFuncName::DPNP_FN_LOG10][eft_INT][eft_INT] = { eft_DBL, (void *)dpnp_log10_c_default}; fmap[DPNPFuncName::DPNP_FN_LOG10][eft_LNG][eft_LNG] = { @@ -780,15 +762,6 @@ static void func_map_init_elemwise_1arg_2type(func_map_t &fmap) fmap[DPNPFuncName::DPNP_FN_TRUNC][eft_DBL][eft_DBL] = { eft_DBL, (void *)dpnp_trunc_c_default}; - fmap[DPNPFuncName::DPNP_FN_TRUNC_EXT][eft_INT][eft_INT] = { - eft_DBL, (void *)dpnp_trunc_c_ext}; - fmap[DPNPFuncName::DPNP_FN_TRUNC_EXT][eft_LNG][eft_LNG] = { - eft_DBL, (void *)dpnp_trunc_c_ext}; - fmap[DPNPFuncName::DPNP_FN_TRUNC_EXT][eft_FLT][eft_FLT] = { - eft_FLT, (void *)dpnp_trunc_c_ext}; - fmap[DPNPFuncName::DPNP_FN_TRUNC_EXT][eft_DBL][eft_DBL] = { - eft_DBL, (void *)dpnp_trunc_c_ext}; - return; } diff --git a/dpnp/dpnp_algo/dpnp_algo.pxd b/dpnp/dpnp_algo/dpnp_algo.pxd index 7427c6c7ce4..80f5471831d 100644 --- a/dpnp/dpnp_algo/dpnp_algo.pxd +++ b/dpnp/dpnp_algo/dpnp_algo.pxd @@ -64,8 +64,6 @@ cdef extern from "dpnp_iface_fptr.hpp" namespace "DPNPFuncName": # need this na DPNP_FN_ASTYPE_EXT DPNP_FN_CBRT DPNP_FN_CBRT_EXT - DPNP_FN_CEIL - DPNP_FN_CEIL_EXT DPNP_FN_CHOLESKY DPNP_FN_CHOLESKY_EXT DPNP_FN_CHOOSE @@ -128,8 +126,6 @@ cdef extern from "dpnp_iface_fptr.hpp" namespace "DPNPFuncName": # need this na DPNP_FN_FILL_DIAGONAL_EXT DPNP_FN_FLATTEN DPNP_FN_FLATTEN_EXT - DPNP_FN_FLOOR - DPNP_FN_FLOOR_EXT DPNP_FN_FMOD DPNP_FN_FMOD_EXT DPNP_FN_FULL @@ -292,8 +288,6 @@ cdef extern from "dpnp_iface_fptr.hpp" namespace "DPNPFuncName": # need this na DPNP_FN_TRIL_EXT DPNP_FN_TRIU DPNP_FN_TRIU_EXT - DPNP_FN_TRUNC - DPNP_FN_TRUNC_EXT DPNP_FN_VANDER DPNP_FN_VANDER_EXT DPNP_FN_VAR diff --git a/dpnp/dpnp_algo/dpnp_algo_mathematical.pxi b/dpnp/dpnp_algo/dpnp_algo_mathematical.pxi index 1685c770d4e..6f5aa2a614c 100644 --- a/dpnp/dpnp_algo/dpnp_algo_mathematical.pxi +++ b/dpnp/dpnp_algo/dpnp_algo_mathematical.pxi @@ -39,7 +39,6 @@ __all__ += [ "dpnp_absolute", "dpnp_arctan2", "dpnp_around", - "dpnp_ceil", "dpnp_conjugate", "dpnp_copysign", "dpnp_cross", @@ -48,7 +47,6 @@ __all__ += [ "dpnp_diff", "dpnp_ediff1d", "dpnp_fabs", - "dpnp_floor", "dpnp_fmod", "dpnp_gradient", 'dpnp_hypot', @@ -65,7 +63,6 @@ __all__ += [ "dpnp_sign", "dpnp_sum", "dpnp_trapz", - "dpnp_trunc" ] @@ -158,10 +155,6 @@ cpdef utils.dpnp_descriptor dpnp_around(utils.dpnp_descriptor x1, int decimals): return result -cpdef utils.dpnp_descriptor dpnp_ceil(utils.dpnp_descriptor x1, utils.dpnp_descriptor out): - return call_fptr_1in_1out_strides(DPNP_FN_CEIL_EXT, x1, dtype=None, out=out, where=True, func_name='ceil') - - cpdef utils.dpnp_descriptor dpnp_conjugate(utils.dpnp_descriptor x1): return call_fptr_1in_1out_strides(DPNP_FN_CONJIGUATE_EXT, x1) @@ -295,10 +288,6 @@ cpdef utils.dpnp_descriptor dpnp_fabs(utils.dpnp_descriptor x1): return call_fptr_1in_1out_strides(DPNP_FN_FABS_EXT, x1) -cpdef utils.dpnp_descriptor dpnp_floor(utils.dpnp_descriptor x1, utils.dpnp_descriptor out): - return call_fptr_1in_1out_strides(DPNP_FN_FLOOR_EXT, x1, dtype=None, out=out, where=True, func_name='floor') - - cpdef utils.dpnp_descriptor dpnp_fmod(utils.dpnp_descriptor x1_obj, utils.dpnp_descriptor x2_obj, object dtype=None, @@ -641,7 +630,3 @@ cpdef utils.dpnp_descriptor dpnp_trapz(utils.dpnp_descriptor y1, utils.dpnp_desc c_dpctl.DPCTLEvent_Delete(event_ref) return result - - -cpdef utils.dpnp_descriptor dpnp_trunc(utils.dpnp_descriptor x1, utils.dpnp_descriptor out): - return call_fptr_1in_1out_strides(DPNP_FN_TRUNC_EXT, x1, dtype=None, out=out, where=True, func_name='trunc') diff --git a/dpnp/dpnp_algo/dpnp_elementwise_common.py b/dpnp/dpnp_algo/dpnp_elementwise_common.py index 33776e26a01..14c5f6058d8 100644 --- a/dpnp/dpnp_algo/dpnp_elementwise_common.py +++ b/dpnp/dpnp_algo/dpnp_elementwise_common.py @@ -46,9 +46,11 @@ "dpnp_bitwise_and", "dpnp_bitwise_or", "dpnp_bitwise_xor", + "dpnp_ceil", "dpnp_cos", "dpnp_divide", "dpnp_equal", + "dpnp_floor", "dpnp_floor_divide", "dpnp_greater", "dpnp_greater_equal", @@ -72,6 +74,7 @@ "dpnp_sqrt", "dpnp_square", "dpnp_subtract", + "dpnp_trunc", ] @@ -318,6 +321,59 @@ def dpnp_bitwise_xor(x1, x2, out=None, order="K"): return dpnp_array._create_from_usm_ndarray(res_usm) +_ceil_docstring = """ +ceil(x, out=None, order='K') + +Returns the ceiling for each element `x_i` for input array `x`. +The ceil of the scalar `x` is the smallest integer `i`, such that `i >= x`. + +Args: + x (dpnp.ndarray): + Input array, expected to have a real-valued data type. + out ({None, dpnp.ndarray}, optional): + Output array to populate. Array must have the correct + shape and the expected data type. + order ("C","F","A","K", optional): memory layout of the new + output array, if parameter `out` is `None`. + Default: "K". +Return: + dpnp.ndarray: + An array containing the element-wise ceiling of input array. + The returned array has the same data type as `x`. +""" + + +def _call_ceil(src, dst, sycl_queue, depends=None): + """A callback to register in UnaryElementwiseFunc class of dpctl.tensor""" + + if depends is None: + depends = [] + + if vmi._mkl_ceil_to_call(sycl_queue, src, dst): + # call pybind11 extension for ceil() function from OneMKL VM + return vmi._ceil(sycl_queue, src, dst, depends) + return ti._ceil(src, dst, sycl_queue, depends) + + +ceil_func = UnaryElementwiseFunc( + "ceil", ti._ceil_result_type, _call_ceil, _ceil_docstring +) + + +def dpnp_ceil(x, out=None, order="K"): + """ + Invokes ceil() function from pybind11 extension of OneMKL VM if possible. + + Otherwise fully relies on dpctl.tensor implementation for ceil() function. + """ + # dpctl.tensor only works with usm_ndarray + x1_usm = dpnp.get_usm_ndarray(x) + out_usm = None if out is None else dpnp.get_usm_ndarray(out) + + res_usm = ceil_func(x1_usm, out=out_usm, order=order) + return dpnp_array._create_from_usm_ndarray(res_usm) + + _cos_docstring = """ cos(x, out=None, order='K') Computes cosine for each element `x_i` for input array `x`. @@ -484,6 +540,59 @@ def dpnp_equal(x1, x2, out=None, order="K"): return dpnp_array._create_from_usm_ndarray(res_usm) +_floor_docstring = """ +floor(x, out=None, order='K') + +Returns the floor for each element `x_i` for input array `x`. +The floor of the scalar `x` is the largest integer `i`, such that `i <= x`. + +Args: + x (dpnp.ndarray): + Input array, expected to have a real-valued data type. + out ({None, dpnp.ndarray}, optional): + Output array to populate. Array must have the correct + shape and the expected data type. + order ("C","F","A","K", optional): memory layout of the new + output array, if parameter `out` is `None`. + Default: "K". +Return: + dpnp.ndarray: + An array containing the element-wise floor of input array. + The returned array has the same data type as `x`. +""" + + +def _call_floor(src, dst, sycl_queue, depends=None): + """A callback to register in UnaryElementwiseFunc class of dpctl.tensor""" + + if depends is None: + depends = [] + + if vmi._mkl_floor_to_call(sycl_queue, src, dst): + # call pybind11 extension for floor() function from OneMKL VM + return vmi._floor(sycl_queue, src, dst, depends) + return ti._floor(src, dst, sycl_queue, depends) + + +floor_func = UnaryElementwiseFunc( + "floor", ti._floor_result_type, _call_floor, _floor_docstring +) + + +def dpnp_floor(x, out=None, order="K"): + """ + Invokes floor() function from pybind11 extension of OneMKL VM if possible. + + Otherwise fully relies on dpctl.tensor implementation for floor() function. + """ + # dpctl.tensor only works with usm_ndarray + x1_usm = dpnp.get_usm_ndarray(x) + out_usm = None if out is None else dpnp.get_usm_ndarray(out) + + res_usm = floor_func(x1_usm, out=out_usm, order=order) + return dpnp_array._create_from_usm_ndarray(res_usm) + + _floor_divide_docstring_ = """ floor_divide(x1, x2, out=None, order="K") Calculates the ratio for each element `x1_i` of the input array `x1` with @@ -1469,3 +1578,58 @@ def dpnp_subtract(x1, x2, out=None, order="K"): ) res_usm = func(x1_usm_or_scalar, x2_usm_or_scalar, out=out_usm, order=order) return dpnp_array._create_from_usm_ndarray(res_usm) + + +_trunc_docstring = """ +trunc(x, out=None, order='K') + +Returns the truncated value for each element `x_i` for input array `x`. +The truncated value of the scalar `x` is the nearest integer `i` which is +closer to zero than `x` is. In short, the fractional part of the +signed number `x` is discarded. + +Args: + x (dpnp.ndarray): + Input array, expected to have a real-valued data type. + out ({None, dpnp.ndarray}, optional): + Output array to populate. Array must have the correct + shape and the expected data type. + order ("C","F","A","K", optional): memory layout of the new + output array, if parameter `out` is `None`. + Default: "K". +Return: + dpnp.ndarray: + An array containing the truncated value of each element in `x`. The data type + of the returned array is determined by the Type Promotion Rules. +""" + + +def _call_trunc(src, dst, sycl_queue, depends=None): + """A callback to register in UnaryElementwiseFunc class of dpctl.tensor""" + + if depends is None: + depends = [] + + if vmi._mkl_trunc_to_call(sycl_queue, src, dst): + # call pybind11 extension for trunc() function from OneMKL VM + return vmi._trunc(sycl_queue, src, dst, depends) + return ti._trunc(src, dst, sycl_queue, depends) + + +trunc_func = UnaryElementwiseFunc( + "trunc", ti._trunc_result_type, _call_trunc, _trunc_docstring +) + + +def dpnp_trunc(x, out=None, order="K"): + """ + Invokes trunc() function from pybind11 extension of OneMKL VM if possible. + + Otherwise fully relies on dpctl.tensor implementation for trunc() function. + """ + # dpctl.tensor only works with usm_ndarray + x1_usm = dpnp.get_usm_ndarray(x) + out_usm = None if out is None else dpnp.get_usm_ndarray(out) + + res_usm = trunc_func(x1_usm, out=out_usm, order=order) + return dpnp_array._create_from_usm_ndarray(res_usm) diff --git a/dpnp/dpnp_iface_mathematical.py b/dpnp/dpnp_iface_mathematical.py index f5c9530046f..8d56d53ca73 100644 --- a/dpnp/dpnp_iface_mathematical.py +++ b/dpnp/dpnp_iface_mathematical.py @@ -51,11 +51,14 @@ from .dpnp_algo.dpnp_elementwise_common import ( check_nd_call_func, dpnp_add, + dpnp_ceil, dpnp_divide, + dpnp_floor, dpnp_floor_divide, dpnp_multiply, dpnp_remainder, dpnp_subtract, + dpnp_trunc, ) from .dpnp_utils import * @@ -145,7 +148,7 @@ def absolute(x, /, out=None, *, where=True, dtype=None, subok=True, **kwargs): ----------- Parameters `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. Parameters `out`, `where`, `dtype` and `subok` are supported with their default values. - Keyword arguments ``kwargs`` are currently unsupported. + Keyword arguments `kwargs` are currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. @@ -270,9 +273,9 @@ def around(x1, decimals=0, out=None): Limitations ----------- - Parameters ``x1`` is supported as :obj:`dpnp.ndarray`. - Parameters ``decimals`` and ``out`` are supported with their default values. - Otherwise the functions will be executed sequentially on CPU. + Parameters `x1` is supported as :class:`dpnp.ndarray`. + Parameters `decimals` and `out` are supported with their default values. + Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. Examples @@ -303,18 +306,34 @@ def around(x1, decimals=0, out=None): return call_origin(numpy.around, x1, decimals=decimals, out=out) -def ceil(x1, out=None, **kwargs): +def ceil( + x, + /, + out=None, + *, + order="K", + where=True, + dtype=None, + subok=True, + **kwargs, +): """ - Compute the ceiling of the input, element-wise. + Compute the ceiling of the input, element-wise. For full documentation refer to :obj:`numpy.ceil`. + Returns + ------- + out : dpnp.ndarray + The ceiling of each element of `x`. + Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. + Parameters `where`, `dtype`, and `subok` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by real-value data types. See Also -------- @@ -325,24 +344,22 @@ def ceil(x1, out=None, **kwargs): -------- >>> import dpnp as np >>> a = np.array([-1.7, -1.5, -0.2, 0.2, 1.5, 1.7, 2.0]) - >>> result = np.ceil(a) - >>> [x for x in result] - [-1.0, -1.0, -0.0, 1.0, 2.0, 2.0, 2.0] + >>> np.ceil(a) + array([-1.0, -1.0, -0.0, 1.0, 2.0, 2.0, 2.0]) """ - x1_desc = dpnp.get_dpnp_descriptor( - x1, copy_when_strides=False, copy_when_nondefault_queue=False + return check_nd_call_func( + numpy.ceil, + dpnp_ceil, + x, + out=out, + where=where, + order=order, + dtype=dtype, + subok=subok, + **kwargs, ) - if x1_desc and not kwargs: - out_desc = ( - dpnp.get_dpnp_descriptor(out, copy_when_nondefault_queue=False) - if out is not None - else None - ) - return dpnp_ceil(x1_desc, out_desc).get_pyobj() - - return call_origin(numpy.ceil, x1, out=out, **kwargs) def conjugate(x1, **kwargs): @@ -404,10 +421,10 @@ def copysign(x1, x2, dtype=None, out=None, where=True, **kwargs): 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. - Otherwise the functions will be executed sequentially on CPU. + Parameters `x1` and `x2` are supported as either :class:`dpnp.ndarray` or scalar. + Parameters `dtype`, `out` and `where` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. Examples @@ -463,12 +480,12 @@ def cross(x1, x2, axisa=-1, axisb=-1, axisc=-1, axis=None): Limitations ----------- - Parameters ``x1`` and ``x2`` are supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Sizes of input arrays are limited by ``x1.size == 3 and x2.size == 3``. - Shapes of input arrays are limited by ``x1.shape == (3,) and x2.shape == (3,)``. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameters `x1` and `x2` are supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Sizes of input arrays are limited by `x1.size == 3 and x2.size == 3`. + Shapes of input arrays are limited by `x1.shape == (3,) and x2.shape == (3,)`. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -511,10 +528,10 @@ def cumprod(x1, **kwargs): Limitations ----------- - Parameter ``x`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -545,10 +562,10 @@ def cumsum(x1, **kwargs): Limitations ----------- - Parameter ``x`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is supported as :obj:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -580,7 +597,7 @@ def diff(x1, n=1, axis=-1, prepend=numpy._NoValue, append=numpy._NoValue): Limitations ----------- Input array is supported as :obj:`dpnp.ndarray`. - Parameters ``axis``, ``prepend`` and ``append`` are supported only with default values. + Parameters `axis`, `prepend` and `append` are supported only with default values. Otherwise the function will be executed sequentially on CPU. """ @@ -633,7 +650,7 @@ def divide( Parameters `x1` and `x2` are supported as either scalar, :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`, but both `x1` and `x2` can not be scalars at the same time. Parameters `where`, `dtype` and `subok` are supported with their default values. - Keyword arguments `kwargs` are currently unsupported. + Keyword argument `kwargs` is currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. @@ -683,12 +700,14 @@ def ediff1d(x1, to_end=None, to_begin=None): Limitations ----------- - Parameter ``x1``is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``to_end`` and ``to_begin`` are currently supported only with default values `None`. - Otherwise the function will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x1`is supported as :class:`dpnp.ndarray`. + Keyword arguments `to_end` and `to_begin` are currently supported only with default values `None`. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. - ..seealso:: :obj:`dpnp.diff` : Calculate the n-th discrete difference along the given axis. + See Also + -------- + :obj:`dpnp.diff` : Calculate the n-th discrete difference along the given axis. Examples -------- @@ -724,12 +743,14 @@ def fabs(x1, **kwargs): Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x1` is supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. - .. seealso:: :obj:`dpnp.abs` : Calculate the absolute value element-wise. + See Also + -------- + :obj:`dpnp.abs` : Calculate the absolute value element-wise. Examples -------- @@ -749,51 +770,65 @@ def fabs(x1, **kwargs): return call_origin(numpy.fabs, x1, **kwargs) -def floor(x1, out=None, **kwargs): +def floor( + x, + /, + out=None, + *, + order="K", + where=True, + dtype=None, + subok=True, + **kwargs, +): """ Round a number to the nearest integer toward minus infinity. For full documentation refer to :obj:`numpy.floor`. + Returns + ------- + out : dpnp.ndarray + The floor of each element of `x`. + Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. + Parameters `where`, `dtype`, and `subok` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by real-value data types. See Also -------- - :obj:`dpnp.ceil` : Compute the ceiling of the input, element-wise. - :obj:`dpnp.trunc` : Return the truncated value of the input, element-wise. + :obj:`dpnp.ceil` : Compute the ceiling of the input, element-wise. + :obj:`dpnp.trunc` : Return the truncated value of the input, element-wise. Notes ----- - Some spreadsheet programs calculate the "floor-towards-zero", in other words floor(-2.5) == -2. - dpNP instead uses the definition of floor where floor(-2.5) == -3. + Some spreadsheet programs calculate the "floor-towards-zero", in other words floor(-2.5) == -2. + DPNP instead uses the definition of floor where floor(-2.5) == -3. Examples -------- >>> import dpnp as np >>> a = np.array([-1.7, -1.5, -0.2, 0.2, 1.5, 1.7, 2.0]) - >>> result = np.floor(a) - >>> [x for x in result] - [-2.0, -2.0, -1.0, 0.0, 1.0, 1.0, 2.0] + >>> np.floor(a) + array([-2.0, -2.0, -1.0, 0.0, 1.0, 1.0, 2.0]) """ - x1_desc = dpnp.get_dpnp_descriptor( - x1, copy_when_strides=False, copy_when_nondefault_queue=False + return check_nd_call_func( + numpy.floor, + dpnp_floor, + x, + out=out, + where=where, + order=order, + dtype=dtype, + subok=subok, + **kwargs, ) - if x1_desc and not kwargs: - out_desc = ( - dpnp.get_dpnp_descriptor(out, copy_when_nondefault_queue=False) - if out is not None - else None - ) - return dpnp_floor(x1_desc, out_desc).get_pyobj() - - return call_origin(numpy.floor, x1, out=out, **kwargs) def floor_divide( @@ -815,11 +850,12 @@ def floor_divide( Limitations ----------- - Parameters ``x1`` and ``x2`` are supported as either :obj:`dpnp.ndarray` or scalar. - Parameters ``where``, ``dtype`` and ``subok`` are supported with their default values. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameters `x1` and `x2` are supported as either scalar, :class:`dpnp.ndarray` + or :class:`dpctl.tensor.usm_ndarray`, but both `x1` and `x2` can not be scalars at the same time. + Parameters `where`, `dtype` and `subok` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. See Also -------- @@ -902,10 +938,10 @@ def fmod(x1, x2, dtype=None, out=None, where=True, **kwargs): 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. - Otherwise the functions will be executed sequentially on CPU. + Parameters `x1` and `x2` are supported as either :class:`dpnp.ndarray` or scalar. + Parameters `dtype`, `out` and `where` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. See Also @@ -973,11 +1009,11 @@ def gradient(x1, *varargs, **kwargs): Limitations ----------- - Parameter ``y1`` is supported as :obj:`dpnp.ndarray`. - Argument ``varargs[0]`` is supported as `int`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `y1` is supported as :class:`dpnp.ndarray`. + Argument `varargs[0]` is supported as `int`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Example ------- @@ -1015,10 +1051,10 @@ def maximum(x1, x2, dtype=None, out=None, where=True, **kwargs): 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. - Otherwise the functions will be executed sequentially on CPU. + Parameters `x1` and `x2` are supported as either :class:`dpnp.ndarray` or scalar. + Parameters `dtype`, `out` and `where` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. See Also @@ -1080,10 +1116,10 @@ def minimum(x1, x2, dtype=None, out=None, where=True, **kwargs): 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. - Otherwise the functions will be executed sequentially on CPU. + Parameters `x1` and `x2` are supported as either :class:`dpnp.ndarray` or scalar. + Parameters `dtype`, `out` and `where` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. See Also @@ -1195,10 +1231,10 @@ def modf(x1, **kwargs): Limitations ----------- - Parameter ``x`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is supported as :obj:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -1245,8 +1281,8 @@ def multiply( Parameters `x1` and `x2` are supported as either scalar, :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`, but both `x1` and `x2` can not be scalars at the same time. Parameters `where`, `dtype` and `subok` are supported with their default values. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. Examples @@ -1294,10 +1330,10 @@ def nancumprod(x1, **kwargs): Limitations ----------- - Parameter ``x`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. .. seealso:: :obj:`dpnp.cumprod` : Return the cumulative product of elements along a given axis. @@ -1331,12 +1367,14 @@ def nancumsum(x1, **kwargs): Limitations ----------- - Parameter ``x`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. - .. seealso:: :obj:`dpnp.cumsum` : Return the cumulative sum of the elements along a given axis. + See Also + -------- + :obj:`dpnp.cumsum` : Return the cumulative sum of the elements along a given axis. Examples -------- @@ -1367,10 +1405,10 @@ def nanprod(x1, **kwargs): Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x1` is supported as :obj:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -1397,10 +1435,10 @@ def nansum(x1, **kwargs): Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x1` is supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -1429,12 +1467,14 @@ def negative(x1, **kwargs): Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x1` is supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. - .. see also: :obj:`dpnp.copysign` : Change the sign of x1 to that of x2, element-wise. + See Also + -------- + :obj:`dpnp.copysign` : Change the sign of x1 to that of x2, element-wise. Examples -------- @@ -1473,7 +1513,7 @@ def power(x1, x2, /, out=None, *, where=True, dtype=None, subok=True, **kwargs): Parameters `x1` and `x2` are supported as either scalar, :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`, but both `x1` and `x2` can not be scalars at the same time. Parameters `where`, `dtype` and `subok` are supported with their default values. - Keyword arguments ``kwargs`` are currently unsupported. + Keyword arguments `kwargs` are currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. @@ -1568,8 +1608,8 @@ def prod( Limitations ----------- - Parameter ``where`` is unsupported. - Input array data types are limited by DPNP :ref:`Data types`. + Parameter `where` is unsupported. + Input array data types are limited by DPNP :ref:`Data types`. Examples -------- @@ -1634,7 +1674,7 @@ def remainder( Parameters `x1` and `x2` are supported as either scalar, :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`, but both `x1` and `x2` can not be scalars at the same time. Parameters `where`, `dtype` and `subok` are supported with their default values. - Keyword arguments `kwargs` are currently unsupported. + Keyword argument `kwargs` is currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. @@ -1685,7 +1725,7 @@ def round_(a, decimals=0, out=None): See Also -------- - :obj:`dpnp.around` : equivalent function; see for details. + :obj:`dpnp.around` : equivalent function; see for details. """ @@ -1700,10 +1740,10 @@ def sign(x1, **kwargs): Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x1` is supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -1749,7 +1789,11 @@ def subtract( ----------- Parameters `x1` and `x2` are supported as either scalar, :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`, but both `x1` and `x2` can not be scalars at the same time. +<<<<<<< HEAD Parameters `where`, `dtype` and `subok` are supported with their default values. +======= + Parameters `out`, `where`, `dtype` and `subok` are supported with their default values. +>>>>>>> use_dpctl_ceil_floor_trunc_in_dpnp Keyword arguments `kwargs` are currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. @@ -1816,11 +1860,11 @@ def sum( Limitations ----------- - Parameters `x` is supported as either :class:`dpnp.ndarray` - or :class:`dpctl.tensor.usm_ndarray`. - Parameters `out`, `initial` and `where` are supported with their default values. - Otherwise the function will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameters `x` is supported as either :class:`dpnp.ndarray` + or :class:`dpctl.tensor.usm_ndarray`. + Parameters `out`, `initial` and `where` are supported with their default values. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -1911,10 +1955,10 @@ def trapz(y1, x1=None, dx=1.0, axis=-1): Limitations ----------- - Parameters ``y`` and ``x`` are supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameters `y` and `x` are supported as :class:`dpnp.ndarray`. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by supported DPNP :ref:`Data types`. Examples -------- @@ -1971,7 +2015,7 @@ def true_divide(*args, **kwargs): See Also -------- - .. seealso:: :obj:`dpnp.divide` : Standard division. + :obj:`dpnp.divide` : Standard division. Notes ----- @@ -1983,43 +2027,57 @@ def true_divide(*args, **kwargs): return dpnp.divide(*args, **kwargs) -def trunc(x1, out=None, **kwargs): +def trunc( + x, + /, + out=None, + *, + order="K", + where=True, + dtype=None, + subok=True, + **kwargs, +): """ Compute the truncated value of the input, element-wise. For full documentation refer to :obj:`numpy.trunc`. + Returns + ------- + out : dpnp.ndarray + The truncated value of each element of `x`. + Limitations ----------- - Parameter ``x1`` is supported as :obj:`dpnp.ndarray`. - Keyword arguments ``kwargs`` are currently unsupported. - Otherwise the functions will be executed sequentially on CPU. - Input array data types are limited by supported DPNP :ref:`Data types`. + Parameter `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. + Parameters `where`, `dtype`, and `subok` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. + Otherwise the function will be executed sequentially on CPU. + Input array data types are limited by real-value data types. See Also -------- - :obj:`dpnp.floor` : Round a number to the nearest integer toward minus infinity. - :obj:`dpnp.ceil` : Round a number to the nearest integer toward infinity. + :obj:`dpnp.floor` : Round a number to the nearest integer toward minus infinity. + :obj:`dpnp.ceil` : Round a number to the nearest integer toward infinity. Examples -------- >>> import dpnp as np >>> a = np.array([-1.7, -1.5, -0.2, 0.2, 1.5, 1.7, 2.0]) - >>> result = np.trunc(a) - >>> [x for x in result] - [-1.0, -1.0, -0.0, 0.0, 1.0, 1.0, 2.0] + >>> np.trunc(a) + array([-1.0, -1.0, -0.0, 0.0, 1.0, 1.0, 2.0]) """ - x1_desc = dpnp.get_dpnp_descriptor( - x1, copy_when_strides=False, copy_when_nondefault_queue=False + return check_nd_call_func( + numpy.trunc, + dpnp_trunc, + x, + out=out, + where=where, + order=order, + dtype=dtype, + subok=subok, + **kwargs, ) - if x1_desc and not kwargs: - out_desc = ( - dpnp.get_dpnp_descriptor(out, copy_when_nondefault_queue=False) - if out is not None - else None - ) - return dpnp_trunc(x1_desc, out_desc).get_pyobj() - - return call_origin(numpy.trunc, x1, out=out, **kwargs) diff --git a/dpnp/dpnp_iface_trigonometric.py b/dpnp/dpnp_iface_trigonometric.py index 904e1e3a1a1..d64359e1db8 100644 --- a/dpnp/dpnp_iface_trigonometric.py +++ b/dpnp/dpnp_iface_trigonometric.py @@ -434,8 +434,9 @@ def cos( Limitations ----------- - Parameters `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. + Parameter `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. Parameters `where`, `dtype` and `subok` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. @@ -986,8 +987,9 @@ def sin( Limitations ----------- - Parameters `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. + Parameter `x` is only supported as either :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`. Parameters `where`, `dtype` and `subok` are supported with their default values. + Keyword arguments `kwargs` are currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. diff --git a/tests/skipped_tests.tbl b/tests/skipped_tests.tbl index 2424ade4cee..c0b051f2d21 100644 --- a/tests/skipped_tests.tbl +++ b/tests/skipped_tests.tbl @@ -806,13 +806,10 @@ tests/third_party/cupy/math_tests/test_rounding.py::TestRoundExtreme_param_5_{de tests/third_party/cupy/math_tests/test_rounding.py::TestRoundExtreme_param_6_{decimals=100}::test_round_large tests/third_party/cupy/math_tests/test_rounding.py::TestRoundExtreme_param_6_{decimals=100}::test_round_small tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_around -tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_ceil tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_fix -tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_floor tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_rint tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_rint_negative tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_round_ -tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_trunc tests/third_party/cupy/math_tests/test_sumprod.py::TestSumprod::test_sum_out tests/third_party/cupy/math_tests/test_sumprod.py::TestSumprod::test_sum_out_wrong_shape tests/third_party/cupy/math_tests/test_sumprod.py::TestCumprod::test_ndarray_cumprod_2dim_with_axis diff --git a/tests/skipped_tests_gpu.tbl b/tests/skipped_tests_gpu.tbl index 85f25b90028..8b05060b707 100644 --- a/tests/skipped_tests_gpu.tbl +++ b/tests/skipped_tests_gpu.tbl @@ -12,23 +12,19 @@ tests/test_random.py::TestPermutationsTestShuffle::test_no_miss_numbers[int64] tests/test_random.py::TestPermutationsTestShuffle::test_shuffle1[lambda x: dpnp.array([])] tests/test_random.py::TestPermutationsTestShuffle::test_shuffle1[lambda x: dpnp.astype(dpnp.asarray(x), dpnp.float32)] -tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-ceil-data1] tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-conjugate-data2] tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-copy-data3] tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-cumprod-data4] tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-cumsum-data5] tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-ediff1d-data7] tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-fabs-data8] -tests/test_sycl_queue.py::test_1in_1out[opencl:gpu:0-floor-data9] -tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-ceil-data1] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-conjugate-data2] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-copy-data3] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-cumprod-data4] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-cumsum-data5] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-ediff1d-data7] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-fabs-data8] -tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-floor-data9] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-nancumprod-data11] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-nancumsum-data12] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-nanprod-data13] @@ -38,7 +34,6 @@ tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-prod-data16] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-sign-data17] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-sum-data18] tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-trapz-data19] -tests/test_sycl_queue.py::test_1in_1out[level_zero:gpu:0-trunc-data20] tests/test_sycl_queue.py::test_modf[level_zero:gpu:0] @@ -992,13 +987,10 @@ tests/third_party/cupy/math_tests/test_rounding.py::TestRoundExtreme_param_5_{de tests/third_party/cupy/math_tests/test_rounding.py::TestRoundExtreme_param_6_{decimals=100}::test_round_large tests/third_party/cupy/math_tests/test_rounding.py::TestRoundExtreme_param_6_{decimals=100}::test_round_small tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_around -tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_ceil tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_fix -tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_floor tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_rint tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_rint_negative tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_round_ -tests/third_party/cupy/math_tests/test_rounding.py::TestRounding::test_trunc tests/third_party/cupy/math_tests/test_sumprod.py::TestCumprod::test_ndarray_cumprod_2dim_with_axis tests/third_party/cupy/math_tests/test_sumprod.py::TestDiff::test_diff_1dim tests/third_party/cupy/math_tests/test_sumprod.py::TestDiff::test_diff_1dim_with_n diff --git a/tests/third_party/cupy/math_tests/test_rounding.py b/tests/third_party/cupy/math_tests/test_rounding.py index e4927495cc5..e17a48aa9e3 100644 --- a/tests/third_party/cupy/math_tests/test_rounding.py +++ b/tests/third_party/cupy/math_tests/test_rounding.py @@ -10,7 +10,7 @@ @testing.gpu class TestRounding(unittest.TestCase): @testing.for_all_dtypes(no_complex=True) - @testing.numpy_cupy_allclose(atol=1e-5) + @testing.numpy_cupy_allclose(type_check=False, atol=1e-5) def check_unary(self, name, xp, dtype): a = testing.shaped_arange((2, 3), xp, dtype) return getattr(xp, name)(a) From 1b069da3a0b23686a3f7c03b473f9555c6ecd67c Mon Sep 17 00:00:00 2001 From: Vahid Tavanashad Date: Mon, 14 Aug 2023 13:16:09 -0500 Subject: [PATCH 2/4] address reviewers' comments --- tests/test_mathematical.py | 91 ++++++++++++++++++++------------------ tests/test_usm_type.py | 8 ++-- 2 files changed, 53 insertions(+), 46 deletions(-) diff --git a/tests/test_mathematical.py b/tests/test_mathematical.py index c5b5ea336ba..5fd00ab726c 100644 --- a/tests/test_mathematical.py +++ b/tests/test_mathematical.py @@ -15,6 +15,7 @@ from .helper import ( get_all_dtypes, get_float_complex_dtypes, + get_float_dtypes, has_support_aspect64, is_cpu_device, is_win_platform, @@ -597,119 +598,125 @@ def test_gradient_y1_dx(self, array, dx): class TestCeil: - def test_ceil(self): + @pytest.mark.parametrize("dtype", get_float_dtypes()) + def test_ceil(self, dtype): array_data = numpy.arange(10) - out = numpy.empty(10, dtype=numpy.float64) + out = numpy.empty(10, dtype) # DPNP - dp_array = dpnp.array(array_data, dtype=dpnp.float64) - dp_out = dpnp.array(out, dtype=dpnp.float64) + dp_array = dpnp.array(array_data, dtype=dtype) + dp_out = dpnp.array(out, dtype=dtype) result = dpnp.ceil(dp_array, out=dp_out) # original - np_array = numpy.array(array_data, dtype=numpy.float64) + np_array = numpy.array(array_data, dtype=dtype) expected = numpy.ceil(np_array, out=out) assert_array_equal(expected, result) @pytest.mark.parametrize( - "dtype", - [numpy.float32, numpy.int64, numpy.int32], - ids=["numpy.float32", "numpy.int64", "numpy.int32"], + "dtype", get_all_dtypes(no_bool=True, no_complex=True, no_none=True) ) def test_invalid_dtype(self, dtype): - dp_array = dpnp.arange(10, dtype=dpnp.float64) + dpnp_dtype = dpnp.float64 if has_support_aspect64() else dpnp.float32 + pytest.skip("similar data types") if dpnp_dtype == dtype else None + dp_array = dpnp.arange(10, dtype=dpnp_dtype) dp_out = dpnp.empty(10, dtype=dtype) - with pytest.raises(ValueError): + with pytest.raises(TypeError): dpnp.ceil(dp_array, out=dp_out) + @pytest.mark.parametrize("dtype", get_float_dtypes()) @pytest.mark.parametrize( "shape", [(0,), (15,), (2, 2)], ids=["(0,)", "(15, )", "(2,2)"] ) - def test_invalid_shape(self, shape): - dp_array = dpnp.arange(10, dtype=dpnp.float64) - dp_out = dpnp.empty(shape, dtype=dpnp.float64) + def test_invalid_shape(self, shape, dtype): + dp_array = dpnp.arange(10, dtype=dtype) + dp_out = dpnp.empty(shape, dtype=dtype) - with pytest.raises(ValueError): + with pytest.raises(TypeError): dpnp.ceil(dp_array, out=dp_out) class TestFloor: - def test_floor(self): + @pytest.mark.parametrize("dtype", get_float_dtypes()) + def test_floor(self, dtype): array_data = numpy.arange(10) - out = numpy.empty(10, dtype=numpy.float64) + out = numpy.empty(10, dtype=dtype) # DPNP - dp_array = dpnp.array(array_data, dtype=dpnp.float64) - dp_out = dpnp.array(out, dtype=dpnp.float64) + dp_array = dpnp.array(array_data, dtype=dtype) + dp_out = dpnp.array(out, dtype=dtype) result = dpnp.floor(dp_array, out=dp_out) # original - np_array = numpy.array(array_data, dtype=numpy.float64) + np_array = numpy.array(array_data, dtype=dtype) expected = numpy.floor(np_array, out=out) assert_array_equal(expected, result) @pytest.mark.parametrize( - "dtype", - [numpy.float32, numpy.int64, numpy.int32], - ids=["numpy.float32", "numpy.int64", "numpy.int32"], + "dtype", get_all_dtypes(no_bool=True, no_complex=True, no_none=True) ) def test_invalid_dtype(self, dtype): - dp_array = dpnp.arange(10, dtype=dpnp.float64) + dpnp_dtype = dpnp.float64 if has_support_aspect64() else dpnp.float32 + pytest.skip("similar data types") if dpnp_dtype == dtype else None + dp_array = dpnp.arange(10, dtype=dpnp_dtype) dp_out = dpnp.empty(10, dtype=dtype) - with pytest.raises(ValueError): + with pytest.raises(TypeError): dpnp.floor(dp_array, out=dp_out) + @pytest.mark.parametrize("dtype", get_float_dtypes()) @pytest.mark.parametrize( "shape", [(0,), (15,), (2, 2)], ids=["(0,)", "(15, )", "(2,2)"] ) - def test_invalid_shape(self, shape): - dp_array = dpnp.arange(10, dtype=dpnp.float64) - dp_out = dpnp.empty(shape, dtype=dpnp.float64) + def test_invalid_shape(self, shape, dtype): + dp_array = dpnp.arange(10, dtype=dtype) + dp_out = dpnp.empty(shape, dtype=dtype) - with pytest.raises(ValueError): + with pytest.raises(TypeError): dpnp.floor(dp_array, out=dp_out) class TestTrunc: - def test_trunc(self): + @pytest.mark.parametrize("dtype", get_float_dtypes()) + def test_trunc(self, dtype): array_data = numpy.arange(10) - out = numpy.empty(10, dtype=numpy.float64) + out = numpy.empty(10, dtype=dtype) # DPNP - dp_array = dpnp.array(array_data, dtype=dpnp.float64) - dp_out = dpnp.array(out, dtype=dpnp.float64) + dp_array = dpnp.array(array_data, dtype=dtype) + dp_out = dpnp.array(out, dtype=dtype) result = dpnp.trunc(dp_array, out=dp_out) # original - np_array = numpy.array(array_data, dtype=numpy.float64) + np_array = numpy.array(array_data, dtype=dtype) expected = numpy.trunc(np_array, out=out) assert_array_equal(expected, result) @pytest.mark.parametrize( - "dtype", - [numpy.float32, numpy.int64, numpy.int32], - ids=["numpy.float32", "numpy.int64", "numpy.int32"], + "dtype", get_all_dtypes(no_bool=True, no_complex=True, no_none=True) ) def test_invalid_dtype(self, dtype): - dp_array = dpnp.arange(10, dtype=dpnp.float64) + dpnp_dtype = dpnp.float64 if has_support_aspect64() else dpnp.float32 + pytest.skip("similar data types") if dpnp_dtype == dtype else None + dp_array = dpnp.arange(10, dtype=dpnp_dtype) dp_out = dpnp.empty(10, dtype=dtype) - with pytest.raises(ValueError): + with pytest.raises(TypeError): dpnp.trunc(dp_array, out=dp_out) + @pytest.mark.parametrize("dtype", get_float_dtypes()) @pytest.mark.parametrize( "shape", [(0,), (15,), (2, 2)], ids=["(0,)", "(15, )", "(2,2)"] ) - def test_invalid_shape(self, shape): - dp_array = dpnp.arange(10, dtype=dpnp.float64) - dp_out = dpnp.empty(shape, dtype=dpnp.float64) + def test_invalid_shape(self, shape, dtype): + dp_array = dpnp.arange(10, dtype=dtype) + dp_out = dpnp.empty(shape, dtype=dtype) - with pytest.raises(ValueError): + with pytest.raises(TypeError): dpnp.trunc(dp_array, out=dp_out) diff --git a/tests/test_usm_type.py b/tests/test_usm_type.py index bda39505a1c..b197897f373 100644 --- a/tests/test_usm_type.py +++ b/tests/test_usm_type.py @@ -230,10 +230,10 @@ def test_meshgrid(usm_type_x, usm_type_y): @pytest.mark.parametrize( "func,data", [ - pytest.param( - "sqrt", - [1.0, 3.0, 9.0], - ), + pytest.param("ceil", [-1.7, -1.5, -0.2, 0.2, 1.5, 1.7, 2.0]), + pytest.param("floor", [-1.7, -1.5, -0.2, 0.2, 1.5, 1.7, 2.0]), + pytest.param("trunc", [-1.7, -1.5, -0.2, 0.2, 1.5, 1.7, 2.0]), + pytest.param("sqrt", [1.0, 3.0, 9.0]), ], ) @pytest.mark.parametrize("usm_type", list_of_usm_types, ids=list_of_usm_types) From 6454b3e8fd1949d8d070f2b6070c96ce591593a2 Mon Sep 17 00:00:00 2001 From: Vahid Tavanashad Date: Wed, 16 Aug 2023 22:16:05 -0500 Subject: [PATCH 3/4] remove skipping tests in gpu_no_f64 --- dpnp/dpnp_iface_mathematical.py | 6 +----- tests/skipped_tests_gpu_no_fp64.tbl | 30 ----------------------------- 2 files changed, 1 insertion(+), 35 deletions(-) diff --git a/dpnp/dpnp_iface_mathematical.py b/dpnp/dpnp_iface_mathematical.py index 8d56d53ca73..724cc6bb643 100644 --- a/dpnp/dpnp_iface_mathematical.py +++ b/dpnp/dpnp_iface_mathematical.py @@ -1789,12 +1789,8 @@ def subtract( ----------- Parameters `x1` and `x2` are supported as either scalar, :class:`dpnp.ndarray` or :class:`dpctl.tensor.usm_ndarray`, but both `x1` and `x2` can not be scalars at the same time. -<<<<<<< HEAD Parameters `where`, `dtype` and `subok` are supported with their default values. -======= - Parameters `out`, `where`, `dtype` and `subok` are supported with their default values. ->>>>>>> use_dpctl_ceil_floor_trunc_in_dpnp - Keyword arguments `kwargs` are currently unsupported. + Keyword argument `kwargs` is currently unsupported. Otherwise the function will be executed sequentially on CPU. Input array data types are limited by supported DPNP :ref:`Data types`. diff --git a/tests/skipped_tests_gpu_no_fp64.tbl b/tests/skipped_tests_gpu_no_fp64.tbl index 2658f178bf8..5d23dad7782 100644 --- a/tests/skipped_tests_gpu_no_fp64.tbl +++ b/tests/skipped_tests_gpu_no_fp64.tbl @@ -350,27 +350,6 @@ tests/test_mathematical.py::TestGradient::test_gradient_y1_dx[2-array2] tests/test_mathematical.py::TestGradient::test_gradient_y1_dx[3.5-array0] tests/test_mathematical.py::TestGradient::test_gradient_y1_dx[3.5-array1] tests/test_mathematical.py::TestGradient::test_gradient_y1_dx[3.5-array2] -tests/test_mathematical.py::TestCeil::test_ceil -tests/test_mathematical.py::TestCeil::test_invalid_dtype[numpy.float32] -tests/test_mathematical.py::TestCeil::test_invalid_dtype[numpy.int64] -tests/test_mathematical.py::TestCeil::test_invalid_dtype[numpy.int32] -tests/test_mathematical.py::TestCeil::test_invalid_shape[(0,)] -tests/test_mathematical.py::TestCeil::test_invalid_shape[(15, )] -tests/test_mathematical.py::TestCeil::test_invalid_shape[(2,2)] -tests/test_mathematical.py::TestFloor::test_floor -tests/test_mathematical.py::TestFloor::test_invalid_dtype[numpy.float32] -tests/test_mathematical.py::TestFloor::test_invalid_dtype[numpy.int64] -tests/test_mathematical.py::TestFloor::test_invalid_dtype[numpy.int32] -tests/test_mathematical.py::TestFloor::test_invalid_shape[(0,)] -tests/test_mathematical.py::TestFloor::test_invalid_shape[(15, )] -tests/test_mathematical.py::TestFloor::test_invalid_shape[(2,2)] -tests/test_mathematical.py::TestTrunc::test_trunc -tests/test_mathematical.py::TestTrunc::test_invalid_dtype[numpy.float32] -tests/test_mathematical.py::TestTrunc::test_invalid_dtype[numpy.int64] -tests/test_mathematical.py::TestTrunc::test_invalid_dtype[numpy.int32] -tests/test_mathematical.py::TestTrunc::test_invalid_shape[(0,)] -tests/test_mathematical.py::TestTrunc::test_invalid_shape[(15, )] -tests/test_mathematical.py::TestTrunc::test_invalid_shape[(2,2)] tests/test_mathematical.py::TestPower::test_power[complex64] tests/test_mathematical.py::TestPower::test_out_dtypes[bool_] tests/test_mathematical.py::TestPower::test_out_dtypes[int32] @@ -397,21 +376,18 @@ tests/test_strides.py::test_strides_1arg[(10,)-int32-arcsinh] tests/test_strides.py::test_strides_1arg[(10,)-int32-arctan] tests/test_strides.py::test_strides_1arg[(10,)-int32-arctanh] tests/test_strides.py::test_strides_1arg[(10,)-int32-cbrt] -tests/test_strides.py::test_strides_1arg[(10,)-int32-ceil] tests/test_strides.py::test_strides_1arg[(10,)-int32-cosh] tests/test_strides.py::test_strides_1arg[(10,)-int32-degrees] tests/test_strides.py::test_strides_1arg[(10,)-int32-exp] tests/test_strides.py::test_strides_1arg[(10,)-int32-exp2] tests/test_strides.py::test_strides_1arg[(10,)-int32-expm1] tests/test_strides.py::test_strides_1arg[(10,)-int32-fabs] -tests/test_strides.py::test_strides_1arg[(10,)-int32-floor] tests/test_strides.py::test_strides_1arg[(10,)-int32-log10] tests/test_strides.py::test_strides_1arg[(10,)-int32-log1p] tests/test_strides.py::test_strides_1arg[(10,)-int32-log2] tests/test_strides.py::test_strides_1arg[(10,)-int32-radians] tests/test_strides.py::test_strides_1arg[(10,)-int32-sinh] tests/test_strides.py::test_strides_1arg[(10,)-int32-tanh] -tests/test_strides.py::test_strides_1arg[(10,)-int32-trunc] tests/test_strides.py::test_strides_1arg[(10,)-int64-arccos] tests/test_strides.py::test_strides_1arg[(10,)-int64-arccosh] tests/test_strides.py::test_strides_1arg[(10,)-int64-arcsin] @@ -419,21 +395,18 @@ tests/test_strides.py::test_strides_1arg[(10,)-int64-arcsinh] tests/test_strides.py::test_strides_1arg[(10,)-int64-arctan] tests/test_strides.py::test_strides_1arg[(10,)-int64-arctanh] tests/test_strides.py::test_strides_1arg[(10,)-int64-cbrt] -tests/test_strides.py::test_strides_1arg[(10,)-int64-ceil] tests/test_strides.py::test_strides_1arg[(10,)-int64-cosh] tests/test_strides.py::test_strides_1arg[(10,)-int64-degrees] tests/test_strides.py::test_strides_1arg[(10,)-int64-exp] tests/test_strides.py::test_strides_1arg[(10,)-int64-exp2] tests/test_strides.py::test_strides_1arg[(10,)-int64-expm1] tests/test_strides.py::test_strides_1arg[(10,)-int64-fabs] -tests/test_strides.py::test_strides_1arg[(10,)-int64-floor] tests/test_strides.py::test_strides_1arg[(10,)-int64-log10] tests/test_strides.py::test_strides_1arg[(10,)-int64-log1p] tests/test_strides.py::test_strides_1arg[(10,)-int64-log2] tests/test_strides.py::test_strides_1arg[(10,)-int64-radians] tests/test_strides.py::test_strides_1arg[(10,)-int64-sinh] tests/test_strides.py::test_strides_1arg[(10,)-int64-tanh] -tests/test_strides.py::test_strides_1arg[(10,)-int64-trunc] tests/test_strides.py::test_strides_1arg[(10,)-None-arccos] tests/test_strides.py::test_strides_1arg[(10,)-None-arccosh] tests/test_strides.py::test_strides_1arg[(10,)-None-arcsin] @@ -441,21 +414,18 @@ tests/test_strides.py::test_strides_1arg[(10,)-None-arcsinh] tests/test_strides.py::test_strides_1arg[(10,)-None-arctan] tests/test_strides.py::test_strides_1arg[(10,)-None-arctanh] tests/test_strides.py::test_strides_1arg[(10,)-None-cbrt] -tests/test_strides.py::test_strides_1arg[(10,)-None-ceil] tests/test_strides.py::test_strides_1arg[(10,)-None-cosh] tests/test_strides.py::test_strides_1arg[(10,)-None-degrees] tests/test_strides.py::test_strides_1arg[(10,)-None-exp] tests/test_strides.py::test_strides_1arg[(10,)-None-exp2] tests/test_strides.py::test_strides_1arg[(10,)-None-expm1] tests/test_strides.py::test_strides_1arg[(10,)-None-fabs] -tests/test_strides.py::test_strides_1arg[(10,)-None-floor] tests/test_strides.py::test_strides_1arg[(10,)-None-log10] tests/test_strides.py::test_strides_1arg[(10,)-None-log1p] tests/test_strides.py::test_strides_1arg[(10,)-None-log2] tests/test_strides.py::test_strides_1arg[(10,)-None-radians] tests/test_strides.py::test_strides_1arg[(10,)-None-sinh] tests/test_strides.py::test_strides_1arg[(10,)-None-tanh] -tests/test_strides.py::test_strides_1arg[(10,)-None-trunc] tests/test_strides.py::test_strides_tan[(10,)-int32] tests/test_strides.py::test_strides_tan[(10,)-int64] tests/test_strides.py::test_strides_tan[(10,)-None] From e0bf7318ae28c9552bbfbea3dae82598139e2ea5 Mon Sep 17 00:00:00 2001 From: Vahid Tavanashad Date: Fri, 18 Aug 2023 07:52:07 -0500 Subject: [PATCH 4/4] modifying related tests suggested by reviewers --- tests/test_mathematical.py | 15 ++++++--------- 1 file changed, 6 insertions(+), 9 deletions(-) diff --git a/tests/test_mathematical.py b/tests/test_mathematical.py index 5fd00ab726c..9d944e3c597 100644 --- a/tests/test_mathematical.py +++ b/tests/test_mathematical.py @@ -615,11 +615,10 @@ def test_ceil(self, dtype): assert_array_equal(expected, result) @pytest.mark.parametrize( - "dtype", get_all_dtypes(no_bool=True, no_complex=True, no_none=True) + "dtype", get_all_dtypes(no_complex=True, no_none=True)[:-1] ) def test_invalid_dtype(self, dtype): - dpnp_dtype = dpnp.float64 if has_support_aspect64() else dpnp.float32 - pytest.skip("similar data types") if dpnp_dtype == dtype else None + dpnp_dtype = get_all_dtypes(no_complex=True, no_none=True)[-1] dp_array = dpnp.arange(10, dtype=dpnp_dtype) dp_out = dpnp.empty(10, dtype=dtype) @@ -656,11 +655,10 @@ def test_floor(self, dtype): assert_array_equal(expected, result) @pytest.mark.parametrize( - "dtype", get_all_dtypes(no_bool=True, no_complex=True, no_none=True) + "dtype", get_all_dtypes(no_complex=True, no_none=True)[:-1] ) def test_invalid_dtype(self, dtype): - dpnp_dtype = dpnp.float64 if has_support_aspect64() else dpnp.float32 - pytest.skip("similar data types") if dpnp_dtype == dtype else None + dpnp_dtype = get_all_dtypes(no_complex=True, no_none=True)[-1] dp_array = dpnp.arange(10, dtype=dpnp_dtype) dp_out = dpnp.empty(10, dtype=dtype) @@ -697,11 +695,10 @@ def test_trunc(self, dtype): assert_array_equal(expected, result) @pytest.mark.parametrize( - "dtype", get_all_dtypes(no_bool=True, no_complex=True, no_none=True) + "dtype", get_all_dtypes(no_complex=True, no_none=True)[:-1] ) def test_invalid_dtype(self, dtype): - dpnp_dtype = dpnp.float64 if has_support_aspect64() else dpnp.float32 - pytest.skip("similar data types") if dpnp_dtype == dtype else None + dpnp_dtype = get_all_dtypes(no_complex=True, no_none=True)[-1] dp_array = dpnp.arange(10, dtype=dpnp_dtype) dp_out = dpnp.empty(10, dtype=dtype)