diff --git a/CMakeLists.txt b/CMakeLists.txt index e797b2f4fe2b..18b543ceb838 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -67,6 +67,62 @@ find_package(Dpctl REQUIRED) message(STATUS "Dpctl_INCLUDE_DIR=" ${Dpctl_INCLUDE_DIR}) message(STATUS "Dpctl_TENSOR_INCLUDE_DIR=" ${Dpctl_TENSOR_INCLUDE_DIR}) +option(DPNP_TARGET_CUDA + "Build DPNP to target CUDA devices" + OFF +) +option(DPNP_USE_ONEMKL_INTERFACES + "Build DPNP with oneMKL Interfaces" + OFF +) +set(_dpnp_sycl_targets) +set(_use_onemkl_interfaces_cuda OFF) +if ("x${DPNP_SYCL_TARGETS}" STREQUAL "x") + if(DPNP_TARGET_CUDA) + set(_dpnp_sycl_targets "nvptx64-nvidia-cuda,spir64-unknown-unknown") + set(_use_onemkl_interfaces_cuda ON) + else() + if(DEFINED ENV{DPNP_TARGET_CUDA}) + set(_dpnp_sycl_targets "nvptx64-nvidia-cuda,spir64-unknown-unknown") + set(_use_onemkl_interfaces_cuda ON) + endif() + endif() +else() + set(_dpnp_sycl_targets ${DPNP_SYCL_TARGETS}) +endif() + +if(_dpnp_sycl_targets) + message(STATUS "Compiling for -fsycl-targets=${_dpnp_sycl_targets}") +endif() + +set(_use_onemkl_interfaces OFF) +if(DPNP_USE_ONEMKL_INTERFACES) + set(_use_onemkl_interfaces ON) +else() + if(DEFINED ENV{DPNP_USE_ONEMKL_INTERFACES}) + set(_use_onemkl_interfaces ON) + endif() +endif() + +if(_use_onemkl_interfaces) + set(BUILD_FUNCTIONAL_TESTS False) + set(BUILD_EXAMPLES False) + if(_use_onemkl_interfaces_cuda) + # set(ENABLE_CUBLAS_BACKEND True) + set(ENABLE_CUSOLVER_BACKEND True) + set(ENABLE_CUFFT_BACKEND True) + # set(ENABLE_CURAND_BACKEND True) + set(ENABLE_MKLGPU_BACKEND False) + set(ENABLE_MKLCPU_BACKEND False) + endif() + FetchContent_Declare( + onemkl_interfaces_library + GIT_REPOSITORY https://github.com/oneapi-src/oneMKL.git + GIT_TAG f2d2dcb4213a435bb60fbb88320c5f24892423ce + ) + FetchContent_MakeAvailable(onemkl_interfaces_library) +endif() + if(WIN32) string(CONCAT WARNING_FLAGS "-Wall " diff --git a/dpnp/backend/extensions/fft/CMakeLists.txt b/dpnp/backend/extensions/fft/CMakeLists.txt index 605fb90acec8..939848a12d69 100644 --- a/dpnp/backend/extensions/fft/CMakeLists.txt +++ b/dpnp/backend/extensions/fft/CMakeLists.txt @@ -34,6 +34,20 @@ set(_module_src pybind11_add_module(${python_module_name} MODULE ${_module_src}) add_sycl_to_target(TARGET ${python_module_name} SOURCES ${_module_src}) +if(_dpnp_sycl_targets) + # make fat binary + target_compile_options( + ${python_module_name} + PRIVATE + -fsycl-targets=${_dpnp_sycl_targets} + ) + target_link_options( + ${python_module_name} + PRIVATE + -fsycl-targets=${_dpnp_sycl_targets} + ) +endif() + if (WIN32) if (${CMAKE_VERSION} VERSION_LESS "3.27") # this is a work-around for target_link_options inserting option after -link option, cause @@ -68,7 +82,12 @@ if (DPNP_GENERATE_COVERAGE) target_link_options(${python_module_name} PRIVATE -fprofile-instr-generate -fcoverage-mapping) endif() -target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::DFT) +if(_use_onemkl_interfaces) + target_link_libraries(${python_module_name} PUBLIC onemkl) + target_compile_options(${python_module_name} PRIVATE -DUSE_ONEMKL_INTERFACES) +else() + target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::DFT) +endif() install(TARGETS ${python_module_name} DESTINATION "dpnp/backend/extensions/fft" diff --git a/dpnp/backend/extensions/fft/common.hpp b/dpnp/backend/extensions/fft/common.hpp index 2affacf09b0d..a9534d95c71f 100644 --- a/dpnp/backend/extensions/fft/common.hpp +++ b/dpnp/backend/extensions/fft/common.hpp @@ -187,26 +187,32 @@ class DescriptorWrapper // config_param::PLACEMENT bool get_in_place() { - // TODO: replace when MKLD-10506 is implemented - // mkl_dft::config_value placement; +#if defined(USE_ONEMKL_INTERFACES) + mkl_dft::config_value placement; + descr_.get_value(mkl_dft::config_param::PLACEMENT, &placement); + return (placement == mkl_dft::config_value::INPLACE); +#else + // TODO: remove branch when MKLD-10506 is implemented DFTI_CONFIG_VALUE placement; - descr_.get_value(mkl_dft::config_param::PLACEMENT, &placement); - // TODO: replace when MKLD-10506 is implemented - // return (placement == mkl_dft::config_value::INPLACE); return (placement == DFTI_CONFIG_VALUE::DFTI_INPLACE); +#endif // USE_ONEMKL_INTERFACES } void set_in_place(const bool &in_place_request) { - // TODO: replace when MKLD-10506 is implemented - // descr_.set_value(mkl_dft::config_param::PLACEMENT, (in_place_request) - // ? mkl_dft::config_value::INPLACE : - // mkl_dft::config_value::NOT_INPLACE); +#if defined(USE_ONEMKL_INTERFACES) + descr_.set_value(mkl_dft::config_param::PLACEMENT, + (in_place_request) + ? mkl_dft::config_value::INPLACE + : mkl_dft::config_value::NOT_INPLACE); +#else + // TODO: remove branch when MKLD-10506 is implemented descr_.set_value(mkl_dft::config_param::PLACEMENT, (in_place_request) ? DFTI_CONFIG_VALUE::DFTI_INPLACE : DFTI_CONFIG_VALUE::DFTI_NOT_INPLACE); +#endif // USE_ONEMKL_INTERFACES } // config_param::PRECISION @@ -221,14 +227,16 @@ class DescriptorWrapper // config_param::COMMIT_STATUS bool is_committed() { - // TODO: replace when MKLD-10506 is implemented - // mkl_dft::config_value committed; +#if defined(USE_ONEMKL_INTERFACES) + mkl_dft::config_value committed; + descr_.get_value(mkl_dft::config_param::COMMIT_STATUS, &committed); + return (committed == mkl_dft::config_value::COMMITTED); +#else + // TODO: remove branch when MKLD-10506 is implemented DFTI_CONFIG_VALUE committed; - descr_.get_value(mkl_dft::config_param::COMMIT_STATUS, &committed); - // TODO: replace when MKLD-10506 is implemented - // return (committed == mkl_dft::config_value::COMMITTED); return (committed == DFTI_CONFIG_VALUE::DFTI_COMMITTED); +#endif // USE_ONEMKL_INTERFACES } private: diff --git a/dpnp/backend/extensions/fft/out_of_place.cpp b/dpnp/backend/extensions/fft/out_of_place.cpp index 70b4164387b0..71474d058e76 100644 --- a/dpnp/backend/extensions/fft/out_of_place.cpp +++ b/dpnp/backend/extensions/fft/out_of_place.cpp @@ -23,7 +23,7 @@ // THE POSSIBILITY OF SUCH DAMAGE. //***************************************************************************** -#include +#include #include #include diff --git a/dpnp/backend/extensions/lapack/CMakeLists.txt b/dpnp/backend/extensions/lapack/CMakeLists.txt index 683bb34009bd..91d2f832d58f 100644 --- a/dpnp/backend/extensions/lapack/CMakeLists.txt +++ b/dpnp/backend/extensions/lapack/CMakeLists.txt @@ -51,6 +51,20 @@ set(_module_src pybind11_add_module(${python_module_name} MODULE ${_module_src}) add_sycl_to_target(TARGET ${python_module_name} SOURCES ${_module_src}) +if(_dpnp_sycl_targets) + # make fat binary + target_compile_options( + ${python_module_name} + PRIVATE + -fsycl-targets=${_dpnp_sycl_targets} + ) + target_link_options( + ${python_module_name} + PRIVATE + -fsycl-targets=${_dpnp_sycl_targets} + ) +endif() + if (WIN32) if (${CMAKE_VERSION} VERSION_LESS "3.27") # this is a work-around for target_link_options inserting option after -link option, cause @@ -85,7 +99,12 @@ if (DPNP_GENERATE_COVERAGE) target_link_options(${python_module_name} PRIVATE -fprofile-instr-generate -fcoverage-mapping) endif() -target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::LAPACK) +if(_use_onemkl_interfaces) + target_link_libraries(${python_module_name} PUBLIC onemkl) + target_compile_options(${python_module_name} PRIVATE -DUSE_ONEMKL_INTERFACES) +else() + target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::LAPACK) +endif() install(TARGETS ${python_module_name} DESTINATION "dpnp/backend/extensions/lapack" diff --git a/dpnp/backend/extensions/lapack/gesv.cpp b/dpnp/backend/extensions/lapack/gesv.cpp index 0cc6eb1f008e..660afb581937 100644 --- a/dpnp/backend/extensions/lapack/gesv.cpp +++ b/dpnp/backend/extensions/lapack/gesv.cpp @@ -56,6 +56,11 @@ static sycl::event gesv_impl(sycl::queue &exec_q, char *in_b, const std::vector &depends) { +#if defined(USE_ONEMKL_INTERFACES) + // Temporary flag for build only + // FIXME: Need to implement by using lapack::getrf and lapack::getrs + std::logic_error("Not Implemented"); +#else type_utils::validate_type_for_device(exec_q); T *a = reinterpret_cast(in_a); @@ -129,6 +134,7 @@ static sycl::event gesv_impl(sycl::queue &exec_q, }); return ht_ev; +#endif // USE_ONEMKL_INTERFACES } std::pair diff --git a/dpnp/backend/extensions/vm/CMakeLists.txt b/dpnp/backend/extensions/vm/CMakeLists.txt index 77f68d35ae83..0bbf6476aa52 100644 --- a/dpnp/backend/extensions/vm/CMakeLists.txt +++ b/dpnp/backend/extensions/vm/CMakeLists.txt @@ -23,47 +23,49 @@ # THE POSSIBILITY OF SUCH DAMAGE. # ***************************************************************************** -set(_elementwise_sources - ${CMAKE_CURRENT_SOURCE_DIR}/abs.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/acos.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/acosh.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/add.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/asin.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/asinh.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/atan.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/atan2.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/atanh.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/cbrt.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/ceil.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/conj.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/cos.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/cosh.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/div.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/exp.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/exp2.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/expm1.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/floor.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/fmax.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/fmin.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/fmod.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/hypot.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/ln.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/log10.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/log1p.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/log2.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/mul.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/nextafter.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/pow.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/rint.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/sin.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/sinh.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/sqr.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/sqrt.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/sub.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/tan.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/tanh.cpp - ${CMAKE_CURRENT_SOURCE_DIR}/trunc.cpp +if(NOT _use_onemkl_interfaces) + set(_elementwise_sources + ${CMAKE_CURRENT_SOURCE_DIR}/abs.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/acos.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/acosh.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/add.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/asin.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/asinh.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/atan.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/atan2.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/atanh.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/cbrt.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/ceil.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/conj.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/cos.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/cosh.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/div.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/exp.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/exp2.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/expm1.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/floor.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/fmax.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/fmin.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/fmod.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/hypot.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/ln.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/log10.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/log1p.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/log2.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/mul.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/nextafter.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/pow.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/rint.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/sin.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/sinh.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/sqr.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/sqrt.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/sub.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/tan.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/tanh.cpp + ${CMAKE_CURRENT_SOURCE_DIR}/trunc.cpp ) +endif() set(_module_src # TODO: remove sources from `elementwise_functions` folder @@ -112,7 +114,11 @@ if (DPNP_GENERATE_COVERAGE) target_link_options(${python_module_name} PRIVATE -fprofile-instr-generate -fcoverage-mapping) endif() -target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::VM) +if(_use_onemkl_interfaces) + target_compile_options(${python_module_name} PRIVATE -DUSE_ONEMKL_INTERFACES) +else() + target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::VM) +endif() install(TARGETS ${python_module_name} DESTINATION "dpnp/backend/extensions/vm" diff --git a/dpnp/backend/extensions/vm/vm_py.cpp b/dpnp/backend/extensions/vm/vm_py.cpp index 72040c1632b7..fb942f6dbaa9 100644 --- a/dpnp/backend/extensions/vm/vm_py.cpp +++ b/dpnp/backend/extensions/vm/vm_py.cpp @@ -23,10 +23,11 @@ // THE POSSIBILITY OF SUCH DAMAGE. //***************************************************************************** // -// This file defines functions of dpnp.backend._lapack_impl extensions +// This file defines functions of dpnp.backend._vm_impl extensions // //***************************************************************************** +#if not defined(USE_ONEMKL_INTERFACES) #include "abs.hpp" #include "acos.hpp" #include "acosh.hpp" @@ -68,9 +69,13 @@ #include "trunc.hpp" namespace vm_ns = dpnp::extensions::vm; +#endif // USE_ONEMKL_INTERFACES + +#include PYBIND11_MODULE(_vm_impl, m) { +#if not defined(USE_ONEMKL_INTERFACES) vm_ns::init_abs(m); vm_ns::init_acos(m); vm_ns::init_acosh(m); @@ -110,4 +115,15 @@ PYBIND11_MODULE(_vm_impl, m) vm_ns::init_tan(m); vm_ns::init_tanh(m); vm_ns::init_trunc(m); +#endif // USE_ONEMKL_INTERFACES + m.def( + "_is_available", + [](void) { +#if defined(USE_ONEMKL_INTERFACES) + return false; +#else + return true; +#endif // USE_ONEMKL_INTERFACES + }, + "Check if the OneMKL VM library can be used."); } diff --git a/dpnp/dpnp_algo/dpnp_elementwise_common.py b/dpnp/dpnp_algo/dpnp_elementwise_common.py index 34a0a630bdf1..2a8e2cc425fc 100644 --- a/dpnp/dpnp_algo/dpnp_elementwise_common.py +++ b/dpnp/dpnp_algo/dpnp_elementwise_common.py @@ -32,6 +32,7 @@ ) import dpnp +import dpnp.backend.extensions.vm._vm_impl as vmi from dpnp.dpnp_array import dpnp_array __all__ = [ @@ -111,11 +112,12 @@ def _call_func(src, dst, sycl_queue, depends=None): if depends is None: depends = [] - if mkl_fn_to_call is not None and mkl_fn_to_call( - sycl_queue, src, dst - ): - # call pybind11 extension for unary function from OneMKL VM - return mkl_impl_fn(sycl_queue, src, dst, depends) + if vmi._is_available() and mkl_fn_to_call is not None: + if getattr(vmi, mkl_fn_to_call)(sycl_queue, src, dst): + # call pybind11 extension for unary function from OneMKL VM + return getattr(vmi, mkl_impl_fn)( + sycl_queue, src, dst, depends + ) return unary_dp_impl_fn(src, dst, sycl_queue, depends) super().__init__( @@ -263,11 +265,12 @@ def _call_func(src1, src2, dst, sycl_queue, depends=None): if depends is None: depends = [] - if mkl_fn_to_call is not None and mkl_fn_to_call( - sycl_queue, src1, src2, dst - ): - # call pybind11 extension for binary function from OneMKL VM - return mkl_impl_fn(sycl_queue, src1, src2, dst, depends) + if vmi._is_available() and mkl_fn_to_call is not None: + if getattr(vmi, mkl_fn_to_call)(sycl_queue, src1, src2, dst): + # call pybind11 extension for binary function from OneMKL VM + return getattr(vmi, mkl_impl_fn)( + sycl_queue, src1, src2, dst, depends + ) return binary_dp_impl_fn(src1, src2, dst, sycl_queue, depends) super().__init__( diff --git a/dpnp/dpnp_iface_mathematical.py b/dpnp/dpnp_iface_mathematical.py index 44eedbb39184..dd384ee2e9e4 100644 --- a/dpnp/dpnp_iface_mathematical.py +++ b/dpnp/dpnp_iface_mathematical.py @@ -55,7 +55,6 @@ import dpnp import dpnp.backend.extensions.ufunc._ufunc_impl as ufi -import dpnp.backend.extensions.vm._vm_impl as vmi from .backend.extensions.sycl_ext import _sycl_ext_impl from .dpnp_algo import ( @@ -369,8 +368,8 @@ def _gradient_num_diff_edges( ti._abs_result_type, ti._abs, _ABS_DOCSTRING, - mkl_fn_to_call=vmi._mkl_abs_to_call, - mkl_impl_fn=vmi._abs, + mkl_fn_to_call="_mkl_abs_to_call", + mkl_impl_fn="_abs", ) @@ -445,8 +444,8 @@ def _gradient_num_diff_edges( ti._add_result_type, ti._add, _ADD_DOCSTRING, - mkl_fn_to_call=vmi._mkl_add_to_call, - mkl_impl_fn=vmi._add, + mkl_fn_to_call="_mkl_add_to_call", + mkl_impl_fn="_add", binary_inplace_fn=ti._add_inplace, ) @@ -593,8 +592,8 @@ def around(x, /, decimals=0, out=None): ti._ceil_result_type, ti._ceil, _CEIL_DOCSTRING, - mkl_fn_to_call=vmi._mkl_ceil_to_call, - mkl_impl_fn=vmi._ceil, + mkl_fn_to_call="_mkl_ceil_to_call", + mkl_impl_fn="_ceil", ) @@ -719,8 +718,8 @@ def clip(a, a_min, a_max, *, out=None, order="K", **kwargs): ti._conj_result_type, ti._conj, _CONJ_DOCSTRING, - mkl_fn_to_call=vmi._mkl_conj_to_call, - mkl_impl_fn=vmi._conj, + mkl_fn_to_call="_mkl_conj_to_call", + mkl_impl_fn="_conj", ) conj = conjugate @@ -1269,8 +1268,8 @@ def diff(a, n=1, axis=-1, prepend=None, append=None): ti._divide_result_type, ti._divide, _DIVIDE_DOCSTRING, - mkl_fn_to_call=vmi._mkl_div_to_call, - mkl_impl_fn=vmi._div, + mkl_fn_to_call="_mkl_div_to_call", + mkl_impl_fn="_div", binary_inplace_fn=ti._divide_inplace, acceptance_fn=_acceptance_fn_divide, ) @@ -1367,8 +1366,8 @@ def ediff1d(x1, to_end=None, to_begin=None): ufi._fabs_result_type, ufi._fabs, _FABS_DOCSTRING, - mkl_fn_to_call=vmi._mkl_abs_to_call, - mkl_impl_fn=vmi._abs, + mkl_fn_to_call="_mkl_abs_to_call", + mkl_impl_fn="_abs", ) @@ -1465,8 +1464,8 @@ def ediff1d(x1, to_end=None, to_begin=None): ufi._float_power_result_type, ti._pow, _FLOAT_POWER_DOCSTRING, - mkl_fn_to_call=vmi._mkl_pow_to_call, - mkl_impl_fn=vmi._pow, + mkl_fn_to_call="_mkl_pow_to_call", + mkl_impl_fn="_pow", binary_inplace_fn=ti._pow_inplace, ) @@ -1524,8 +1523,8 @@ def ediff1d(x1, to_end=None, to_begin=None): ti._floor_result_type, ti._floor, _FLOOR_DOCSTRING, - mkl_fn_to_call=vmi._mkl_floor_to_call, - mkl_impl_fn=vmi._floor, + mkl_fn_to_call="_mkl_floor_to_call", + mkl_impl_fn="_floor", ) @@ -1678,8 +1677,8 @@ def ediff1d(x1, to_end=None, to_begin=None): ufi._fmax_result_type, ufi._fmax, _FMAX_DOCSTRING, - mkl_fn_to_call=vmi._mkl_fmax_to_call, - mkl_impl_fn=vmi._fmax, + mkl_fn_to_call="_mkl_fmax_to_call", + mkl_impl_fn="_fmax", ) @@ -1763,8 +1762,8 @@ def ediff1d(x1, to_end=None, to_begin=None): ufi._fmin_result_type, ufi._fmin, _FMIN_DOCSTRING, - mkl_fn_to_call=vmi._mkl_fmin_to_call, - mkl_impl_fn=vmi._fmin, + mkl_fn_to_call="_mkl_fmin_to_call", + mkl_impl_fn="_fmin", ) @@ -1837,8 +1836,8 @@ def ediff1d(x1, to_end=None, to_begin=None): ufi._fmod_result_type, ufi._fmod, _FMOD_DOCSTRING, - mkl_fn_to_call=vmi._mkl_fmod_to_call, - mkl_impl_fn=vmi._fmod, + mkl_fn_to_call="_mkl_fmod_to_call", + mkl_impl_fn="_fmod", ) @@ -2355,8 +2354,8 @@ def modf(x1, **kwargs): ti._multiply_result_type, ti._multiply, _MULTIPLY_DOCSTRING, - mkl_fn_to_call=vmi._mkl_mul_to_call, - mkl_impl_fn=vmi._mul, + mkl_fn_to_call="_mkl_mul_to_call", + mkl_impl_fn="_mul", binary_inplace_fn=ti._multiply_inplace, ) @@ -2605,8 +2604,8 @@ def nan_to_num(x, copy=True, nan=0.0, posinf=None, neginf=None): ti._nextafter_result_type, ti._nextafter, _NEXTAFTER_DOCSTRING, - mkl_fn_to_call=vmi._mkl_nextafter_to_call, - mkl_impl_fn=vmi._nextafter, + mkl_fn_to_call="_mkl_nextafter_to_call", + mkl_impl_fn="_nextafter", ) @@ -2752,8 +2751,8 @@ def nan_to_num(x, copy=True, nan=0.0, posinf=None, neginf=None): ti._pow_result_type, ti._pow, _POWER_DOCSTRING, - mkl_fn_to_call=vmi._mkl_pow_to_call, - mkl_impl_fn=vmi._pow, + mkl_fn_to_call="_mkl_pow_to_call", + mkl_impl_fn="_pow", binary_inplace_fn=ti._pow_inplace, ) @@ -3067,8 +3066,8 @@ def prod( ti._round_result_type, ti._round, _RINT_DOCSTRING, - mkl_fn_to_call=vmi._mkl_round_to_call, - mkl_impl_fn=vmi._round, + mkl_fn_to_call="_mkl_round_to_call", + mkl_impl_fn="_round", ) @@ -3127,8 +3126,8 @@ def prod( ti._round_result_type, ti._round, _ROUND_DOCSTRING, - mkl_fn_to_call=vmi._mkl_round_to_call, - mkl_impl_fn=vmi._round, + mkl_fn_to_call="_mkl_round_to_call", + mkl_impl_fn="_round", ) @@ -3307,8 +3306,8 @@ def prod( ti._subtract_result_type, ti._subtract, _SUBTRACT_DOCSTRING, - mkl_fn_to_call=vmi._mkl_sub_to_call, - mkl_impl_fn=vmi._sub, + mkl_fn_to_call="_mkl_sub_to_call", + mkl_impl_fn="_sub", binary_inplace_fn=ti._subtract_inplace, acceptance_fn=acceptance_fn_subtract, ) @@ -3551,6 +3550,6 @@ def trapz(y1, x1=None, dx=1.0, axis=-1): ti._trunc_result_type, ti._trunc, _TRUNC_DOCSTRING, - mkl_fn_to_call=vmi._mkl_trunc_to_call, - mkl_impl_fn=vmi._trunc, + mkl_fn_to_call="_mkl_trunc_to_call", + mkl_impl_fn="_trunc", ) diff --git a/dpnp/dpnp_iface_trigonometric.py b/dpnp/dpnp_iface_trigonometric.py index fbfa035f122b..68b9fcaf13a0 100644 --- a/dpnp/dpnp_iface_trigonometric.py +++ b/dpnp/dpnp_iface_trigonometric.py @@ -47,7 +47,6 @@ import dpnp import dpnp.backend.extensions.ufunc._ufunc_impl as ufi -import dpnp.backend.extensions.vm._vm_impl as vmi from .dpnp_algo.dpnp_elementwise_common import DPNPBinaryFunc, DPNPUnaryFunc from .dpnp_utils.dpnp_utils_reduction import dpnp_wrap_reduction_call @@ -152,8 +151,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._acos_result_type, ti._acos, _ACOS_DOCSTRING, - mkl_fn_to_call=vmi._mkl_acos_to_call, - mkl_impl_fn=vmi._acos, + mkl_fn_to_call="_mkl_acos_to_call", + mkl_impl_fn="_acos", ) @@ -207,8 +206,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._acosh_result_type, ti._acosh, _ACOSH_DOCSTRING, - mkl_fn_to_call=vmi._mkl_acosh_to_call, - mkl_impl_fn=vmi._acosh, + mkl_fn_to_call="_mkl_acosh_to_call", + mkl_impl_fn="_acosh", ) @@ -262,8 +261,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._asin_result_type, ti._asin, _ASIN_DOCSTRING, - mkl_fn_to_call=vmi._mkl_asin_to_call, - mkl_impl_fn=vmi._asin, + mkl_fn_to_call="_mkl_asin_to_call", + mkl_impl_fn="_asin", ) @@ -317,8 +316,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._asinh_result_type, ti._asinh, _ASINH_DOCSTRING, - mkl_fn_to_call=vmi._mkl_asinh_to_call, - mkl_impl_fn=vmi._asinh, + mkl_fn_to_call="_mkl_asinh_to_call", + mkl_impl_fn="_asinh", ) @@ -374,8 +373,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._atan_result_type, ti._atan, _ATAN_DOCSTRING, - mkl_fn_to_call=vmi._mkl_atan_to_call, - mkl_impl_fn=vmi._atan, + mkl_fn_to_call="_mkl_atan_to_call", + mkl_impl_fn="_atan", ) @@ -450,8 +449,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._atan2_result_type, ti._atan2, _ATAN2_DOCSTRING, - mkl_fn_to_call=vmi._mkl_atan2_to_call, - mkl_impl_fn=vmi._atan2, + mkl_fn_to_call="_mkl_atan2_to_call", + mkl_impl_fn="_atan2", ) @@ -505,8 +504,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._atanh_result_type, ti._atanh, _ATANH_DOCSTRING, - mkl_fn_to_call=vmi._mkl_atanh_to_call, - mkl_impl_fn=vmi._atanh, + mkl_fn_to_call="_mkl_atanh_to_call", + mkl_impl_fn="_atanh", ) @@ -557,8 +556,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._cbrt_result_type, ti._cbrt, _CBRT_DOCSTRING, - mkl_fn_to_call=vmi._mkl_cbrt_to_call, - mkl_impl_fn=vmi._cbrt, + mkl_fn_to_call="_mkl_cbrt_to_call", + mkl_impl_fn="_cbrt", ) @@ -611,8 +610,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._cos_result_type, ti._cos, _COS_DOCSTRING, - mkl_fn_to_call=vmi._mkl_cos_to_call, - mkl_impl_fn=vmi._cos, + mkl_fn_to_call="_mkl_cos_to_call", + mkl_impl_fn="_cos", ) @@ -666,8 +665,8 @@ def _get_accumulation_res_dt(a, dtype, _out): ti._cosh_result_type, ti._cosh, _COSH_DOCSTRING, - mkl_fn_to_call=vmi._mkl_cosh_to_call, - mkl_impl_fn=vmi._cosh, + mkl_fn_to_call="_mkl_cosh_to_call", + mkl_impl_fn="_cosh", ) @@ -920,8 +919,8 @@ def cumlogsumexp( ti._exp_result_type, ti._exp, _EXP_DOCSTRING, - mkl_fn_to_call=vmi._mkl_exp_to_call, - mkl_impl_fn=vmi._exp, + mkl_fn_to_call="_mkl_exp_to_call", + mkl_impl_fn="_exp", ) @@ -974,8 +973,8 @@ def cumlogsumexp( ti._exp2_result_type, ti._exp2, _EXP2_DOCSTRING, - mkl_fn_to_call=vmi._mkl_exp2_to_call, - mkl_impl_fn=vmi._exp2, + mkl_fn_to_call="_mkl_exp2_to_call", + mkl_impl_fn="_exp2", ) @@ -1036,8 +1035,8 @@ def cumlogsumexp( ti._expm1_result_type, ti._expm1, _EXPM1_DOCSTRING, - mkl_fn_to_call=vmi._mkl_expm1_to_call, - mkl_impl_fn=vmi._expm1, + mkl_fn_to_call="_mkl_expm1_to_call", + mkl_impl_fn="_expm1", ) @@ -1102,8 +1101,8 @@ def cumlogsumexp( ti._hypot_result_type, ti._hypot, _HYPOT_DOCSTRING, - mkl_fn_to_call=vmi._mkl_hypot_to_call, - mkl_impl_fn=vmi._hypot, + mkl_fn_to_call="_mkl_hypot_to_call", + mkl_impl_fn="_hypot", ) @@ -1157,8 +1156,8 @@ def cumlogsumexp( ti._log_result_type, ti._log, _LOG_DOCSTRING, - mkl_fn_to_call=vmi._mkl_ln_to_call, - mkl_impl_fn=vmi._ln, + mkl_fn_to_call="_mkl_ln_to_call", + mkl_impl_fn="_ln", ) @@ -1214,8 +1213,8 @@ def cumlogsumexp( ti._log10_result_type, ti._log10, _LOG10_DOCSTRING, - mkl_fn_to_call=vmi._mkl_log10_to_call, - mkl_impl_fn=vmi._log10, + mkl_fn_to_call="_mkl_log10_to_call", + mkl_impl_fn="_log10", ) @@ -1277,8 +1276,8 @@ def cumlogsumexp( ti._log1p_result_type, ti._log1p, _LOG1P_DOCSTRING, - mkl_fn_to_call=vmi._mkl_log1p_to_call, - mkl_impl_fn=vmi._log1p, + mkl_fn_to_call="_mkl_log1p_to_call", + mkl_impl_fn="_log1p", ) @@ -1335,8 +1334,8 @@ def cumlogsumexp( ti._log2_result_type, ti._log2, _LOG2_DOCSTRING, - mkl_fn_to_call=vmi._mkl_log2_to_call, - mkl_impl_fn=vmi._log2, + mkl_fn_to_call="_mkl_log2_to_call", + mkl_impl_fn="_log2", ) @@ -1910,8 +1909,8 @@ def reduce_hypot(x, /, *, axis=None, dtype=None, keepdims=False, out=None): ti._sin_result_type, ti._sin, _SIN_DOCSTRING, - mkl_fn_to_call=vmi._mkl_sin_to_call, - mkl_impl_fn=vmi._sin, + mkl_fn_to_call="_mkl_sin_to_call", + mkl_impl_fn="_sin", ) @@ -1963,8 +1962,8 @@ def reduce_hypot(x, /, *, axis=None, dtype=None, keepdims=False, out=None): ti._sinh_result_type, ti._sinh, _SINH_DOCSTRING, - mkl_fn_to_call=vmi._mkl_sinh_to_call, - mkl_impl_fn=vmi._sinh, + mkl_fn_to_call="_mkl_sinh_to_call", + mkl_impl_fn="_sinh", ) @@ -2019,8 +2018,8 @@ def reduce_hypot(x, /, *, axis=None, dtype=None, keepdims=False, out=None): ti._sqrt_result_type, ti._sqrt, _SQRT_DOCSTRING, - mkl_fn_to_call=vmi._mkl_sqrt_to_call, - mkl_impl_fn=vmi._sqrt, + mkl_fn_to_call="_mkl_sqrt_to_call", + mkl_impl_fn="_sqrt", ) @@ -2074,8 +2073,8 @@ def reduce_hypot(x, /, *, axis=None, dtype=None, keepdims=False, out=None): ti._square_result_type, ti._square, _SQUARE_DOCSTRING, - mkl_fn_to_call=vmi._mkl_sqr_to_call, - mkl_impl_fn=vmi._sqr, + mkl_fn_to_call="_mkl_sqr_to_call", + mkl_impl_fn="_sqr", ) @@ -2128,8 +2127,8 @@ def reduce_hypot(x, /, *, axis=None, dtype=None, keepdims=False, out=None): ti._tan_result_type, ti._tan, _TAN_DOCSTRING, - mkl_fn_to_call=vmi._mkl_tan_to_call, - mkl_impl_fn=vmi._tan, + mkl_fn_to_call="_mkl_tan_to_call", + mkl_impl_fn="_tan", ) @@ -2182,8 +2181,8 @@ def reduce_hypot(x, /, *, axis=None, dtype=None, keepdims=False, out=None): ti._tanh_result_type, ti._tanh, _TANH_DOCSTRING, - mkl_fn_to_call=vmi._mkl_tanh_to_call, - mkl_impl_fn=vmi._tanh, + mkl_fn_to_call="_mkl_tanh_to_call", + mkl_impl_fn="_tanh", ) diff --git a/scripts/build_locally.py b/scripts/build_locally.py index a0e5c55edc98..d5f14102837e 100644 --- a/scripts/build_locally.py +++ b/scripts/build_locally.py @@ -38,6 +38,8 @@ def run( cmake_executable=None, verbose=False, cmake_opts="", + target="intel", + onemkl_interfaces=False, ): build_system = None @@ -93,6 +95,19 @@ def run( if use_oneapi: if "DPL_ROOT" in os.environ: os.environ["DPL_ROOT_HINT"] = os.environ["DPL_ROOT"] + + if target == "cuda": + cmake_args += [ + "-DDPNP_TARGET_CUDA=ON", + ] + # Always builds using oneMKL interfaces for the cuda target + onemkl_interfaces = True + + if onemkl_interfaces: + cmake_args += [ + "-DDPNP_USE_ONEMKL_INTERFACES=ON", + ] + subprocess.check_call( cmake_args, shell=False, cwd=setup_dir, env=os.environ ) @@ -147,6 +162,19 @@ def run( default="", type=str, ) + driver.add_argument( + "--target", + help="Target backend for build", + dest="target", + default="intel", + type=str, + ) + driver.add_argument( + "--onemkl_interfaces", + help="Build using oneMKL Interfaces", + dest="onemkl_interfaces", + action="store_true", + ) args = parser.parse_args() args_to_validate = [ @@ -200,4 +228,6 @@ def run( cmake_executable=args.cmake_executable, verbose=args.verbose, cmake_opts=args.cmake_opts, + target=args.target, + onemkl_interfaces=args.onemkl_interfaces, )