Skip to content

Commit 42f6815

Browse files
antonwolfyantonwolfy
authored
Merge branch 'master' into bump-test-basic
2 parents 5b19ed7 + acb74b9 commit 42f6815

32 files changed

+1790
-447
lines changed

‎CMakeLists.txt

Lines changed: 3 additions & 10 deletions
Original file line numberDiff line numberDiff line change
@@ -26,18 +26,11 @@ endif()
2626
set(MKL_ARCH "intel64")
2727
set(MKL_LINK "dynamic")
2828
set(MKL_THREADING "tbb_thread")
29-
set(MKL_VERSION_2024 FALSE)
29+
set(MKL_INTERFACE "ilp64")
3030
find_package(MKL QUIET)
3131
if(MKL_FOUND)
32-
if(MKL_VERSION VERSION_GREATER_EQUAL "2024.0.0")
33-
set(MKL_VERSION_2024 TRUE)
34-
set(MKL_INTERFACE "ilp64")
35-
find_package(MKL REQUIRED)
36-
endif()
37-
endif()
38-
39-
if(NOT MKL_VERSION_2024)
40-
set(MKL_INTERFACE_FULL "intel_ilp64")
32+
find_package(MKL REQUIRED)
33+
else()
4134
find_package(MKL REQUIRED PATHS ${CMAKE_SOURCE_DIR}/dpnp/backend/cmake/Modules NO_DEFAULT_PATH)
4235
endif()
4336

‎doc/reference/logic.rst

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -26,6 +26,8 @@ Infinities and NaNs
2626
dpnp.isfinite
2727
dpnp.isinf
2828
dpnp.isnan
29+
dpnp.isneginf
30+
dpnp.isposinf
2931

3032

3133
Array type testing

‎dpnp/backend/CMakeLists.txt

Lines changed: 1 addition & 6 deletions
Original file line numberDiff line numberDiff line change
@@ -84,12 +84,7 @@ if(DPNP_GENERATE_COVERAGE)
8484
target_link_options(${_trgt} PRIVATE -fprofile-instr-generate -fcoverage-mapping)
8585
endif()
8686

87-
if (MKL_VERSION_2024)
88-
target_link_libraries(${_trgt} PUBLIC MKL::MKL_SYCL)
89-
else()
90-
target_link_libraries(${_trgt} PUBLIC MKL::MKL_DPCPP)
91-
endif()
92-
87+
target_link_libraries(${_trgt} PUBLIC MKL::MKL_SYCL)
9388
target_link_libraries(${_trgt} PUBLIC oneDPL)
9489

9590
if (UNIX)

‎dpnp/backend/extensions/blas/CMakeLists.txt

Lines changed: 1 addition & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -69,11 +69,7 @@ if (DPNP_GENERATE_COVERAGE)
6969
target_link_options(${python_module_name} PRIVATE -fprofile-instr-generate -fcoverage-mapping)
7070
endif()
7171

72-
if (MKL_VERSION_2024)
73-
target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::BLAS)
74-
else()
75-
target_link_libraries(${python_module_name} PUBLIC MKL::MKL_DPCPP)
76-
endif()
72+
target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::BLAS)
7773

7874
install(TARGETS ${python_module_name}
7975
DESTINATION "dpnp/backend/extensions/blas"

‎dpnp/backend/extensions/lapack/CMakeLists.txt

Lines changed: 1 addition & 5 deletions
Original file line numberDiff line numberDiff line change
@@ -82,11 +82,7 @@ if (DPNP_GENERATE_COVERAGE)
8282
target_link_options(${python_module_name} PRIVATE -fprofile-instr-generate -fcoverage-mapping)
8383
endif()
8484

85-
if (MKL_VERSION_2024)
86-
target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::LAPACK)
87-
else()
88-
target_link_libraries(${python_module_name} PUBLIC MKL::MKL_DPCPP)
89-
endif()
85+
target_link_libraries(${python_module_name} PUBLIC MKL::MKL_SYCL::LAPACK)
9086

9187
install(TARGETS ${python_module_name}
9288
DESTINATION "dpnp/backend/extensions/lapack"

‎dpnp/backend/extensions/ufunc/CMakeLists.txt

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -26,6 +26,7 @@
2626
set(_elementwise_sources
2727
${CMAKE_CURRENT_SOURCE_DIR}/elementwise_functions/common.cpp
2828
${CMAKE_CURRENT_SOURCE_DIR}/elementwise_functions/fabs.cpp
29+
${CMAKE_CURRENT_SOURCE_DIR}/elementwise_functions/fmod.cpp
2930
)
3031

3132
set(python_module_name _ufunc_impl)

‎dpnp/backend/extensions/ufunc/elementwise_functions/common.cpp

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -26,6 +26,7 @@
2626
#include <pybind11/pybind11.h>
2727

2828
#include "fabs.hpp"
29+
#include "fmod.hpp"
2930

3031
namespace py = pybind11;
3132

@@ -37,5 +38,6 @@ namespace dpnp::extensions::ufunc
3738
void init_elementwise_functions(py::module_ m)
3839
{
3940
init_fabs(m);
41+
init_fmod(m);
4042
}
4143
} // namespace dpnp::extensions::ufunc

‎dpnp/backend/extensions/ufunc/elementwise_functions/fmod.cpp

Lines changed: 193 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,193 @@
1+
//*****************************************************************************
2+
// Copyright (c) 2024, Intel Corporation
3+
// All rights reserved.
4+
//
5+
// Redistribution and use in source and binary forms, with or without
6+
// modification, are permitted provided that the following conditions are met:
7+
// - Redistributions of source code must retain the above copyright notice,
8+
// this list of conditions and the following disclaimer.
9+
// - Redistributions in binary form must reproduce the above copyright notice,
10+
// this list of conditions and the following disclaimer in the documentation
11+
// and/or other materials provided with the distribution.
12+
//
13+
// THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS"
14+
// AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
15+
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE
16+
// ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE
17+
// LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR
18+
// CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF
19+
// SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS
20+
// INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN
21+
// CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE)
22+
// ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF
23+
// THE POSSIBILITY OF SUCH DAMAGE.
24+
//*****************************************************************************
25+
26+
#include <sycl/sycl.hpp>
27+
28+
#include "dpctl4pybind11.hpp"
29+
30+
#include "fmod.hpp"
31+
#include "kernels/elementwise_functions/fmod.hpp"
32+
#include "populate.hpp"
33+
34+
// include a local copy of elementwise common header from dpctl tensor:
35+
// dpctl/tensor/libtensor/source/elementwise_functions/elementwise_functions.hpp
36+
// TODO: replace by including dpctl header once available
37+
#include "../../elementwise_functions/elementwise_functions.hpp"
38+
39+
// dpctl tensor headers
40+
#include "kernels/elementwise_functions/common.hpp"
41+
#include "utils/type_dispatch.hpp"
42+
43+
namespace py = pybind11;
44+
45+
namespace dpnp::extensions::ufunc
46+
{
47+
namespace ew_cmn_ns = dpctl::tensor::kernels::elementwise_common;
48+
namespace py_int = dpnp::extensions::py_internal;
49+
namespace td_ns = dpctl::tensor::type_dispatch;
50+
51+
using ew_cmn_ns::unary_contig_impl_fn_ptr_t;
52+
using ew_cmn_ns::unary_strided_impl_fn_ptr_t;
53+
54+
namespace impl
55+
{
56+
/**
57+
* @brief A factory to define pairs of supported types for which
58+
* sycl::fmod<T> function is available.
59+
*
60+
* @tparam T1 Type of input vectors `a`
61+
* @tparam T2 Type of input vectors `b`
62+
*/
63+
template <typename T1, typename T2>
64+
struct OutputType
65+
{
66+
using value_type = typename std::disjunction<
67+
td_ns::BinaryTypeMapResultEntry<T1, bool, T2, bool, std::int8_t>,
68+
td_ns::BinaryTypeMapResultEntry<T1,
69+
std::uint8_t,
70+
T2,
71+
std::uint8_t,
72+
std::uint8_t>,
73+
td_ns::BinaryTypeMapResultEntry<T1,
74+
std::int8_t,
75+
T2,
76+
std::int8_t,
77+
std::int8_t>,
78+
td_ns::BinaryTypeMapResultEntry<T1,
79+
std::uint16_t,
80+
T2,
81+
std::uint16_t,
82+
std::uint16_t>,
83+
td_ns::BinaryTypeMapResultEntry<T1,
84+
std::int16_t,
85+
T2,
86+
std::int16_t,
87+
std::int16_t>,
88+
td_ns::BinaryTypeMapResultEntry<T1,
89+
std::uint32_t,
90+
T2,
91+
std::uint32_t,
92+
std::uint32_t>,
93+
td_ns::BinaryTypeMapResultEntry<T1,
94+
std::int32_t,
95+
T2,
96+
std::int32_t,
97+
std::int32_t>,
98+
td_ns::BinaryTypeMapResultEntry<T1,
99+
std::uint64_t,
100+
T2,
101+
std::uint64_t,
102+
std::uint64_t>,
103+
td_ns::BinaryTypeMapResultEntry<T1,
104+
std::int64_t,
105+
T2,
106+
std::int64_t,
107+
std::int64_t>,
108+
td_ns::BinaryTypeMapResultEntry<T1,
109+
sycl::half,
110+
T2,
111+
sycl::half,
112+
sycl::half>,
113+
td_ns::BinaryTypeMapResultEntry<T1, float, T2, float, float>,
114+
td_ns::BinaryTypeMapResultEntry<T1, double, T2, double, double>,
115+
td_ns::DefaultResultEntry<void>>::result_type;
116+
};
117+
118+
using dpnp::kernels::fmod::FmodFunctor;
119+
120+
template <typename argT1,
121+
typename argT2,
122+
typename resT,
123+
unsigned int vec_sz = 4,
124+
unsigned int n_vecs = 2,
125+
bool enable_sg_loadstore = true>
126+
using ContigFunctor =
127+
ew_cmn_ns::BinaryContigFunctor<argT1,
128+
argT2,
129+
resT,
130+
FmodFunctor<argT1, argT2, resT>,
131+
vec_sz,
132+
n_vecs,
133+
enable_sg_loadstore>;
134+
135+
template <typename argT1, typename argT2, typename resT, typename IndexerT>
136+
using StridedFunctor =
137+
ew_cmn_ns::BinaryStridedFunctor<argT1,
138+
argT2,
139+
resT,
140+
IndexerT,
141+
FmodFunctor<argT1, argT2, resT>>;
142+
143+
using ew_cmn_ns::binary_contig_impl_fn_ptr_t;
144+
using ew_cmn_ns::binary_contig_matrix_contig_row_broadcast_impl_fn_ptr_t;
145+
using ew_cmn_ns::binary_contig_row_contig_matrix_broadcast_impl_fn_ptr_t;
146+
using ew_cmn_ns::binary_strided_impl_fn_ptr_t;
147+
148+
static binary_contig_impl_fn_ptr_t fmod_contig_dispatch_table[td_ns::num_types]
149+
[td_ns::num_types];
150+
static int fmod_output_typeid_table[td_ns::num_types][td_ns::num_types];
151+
static binary_strided_impl_fn_ptr_t
152+
fmod_strided_dispatch_table[td_ns::num_types][td_ns::num_types];
153+
154+
MACRO_POPULATE_DISPATCH_TABLES(fmod);
155+
} // namespace impl
156+
157+
void init_fmod(py::module_ m)
158+
{
159+
using arrayT = dpctl::tensor::usm_ndarray;
160+
using event_vecT = std::vector<sycl::event>;
161+
{
162+
impl::populate_fmod_dispatch_tables();
163+
using impl::fmod_contig_dispatch_table;
164+
using impl::fmod_output_typeid_table;
165+
using impl::fmod_strided_dispatch_table;
166+
167+
auto fmod_pyapi = [&](const arrayT &src1, const arrayT &src2,
168+
const arrayT &dst, sycl::queue &exec_q,
169+
const event_vecT &depends = {}) {
170+
return py_int::py_binary_ufunc(
171+
src1, src2, dst, exec_q, depends, fmod_output_typeid_table,
172+
fmod_contig_dispatch_table, fmod_strided_dispatch_table,
173+
// no support of C-contig row with broadcasting in OneMKL
174+
td_ns::NullPtrTable<
175+
impl::
176+
binary_contig_matrix_contig_row_broadcast_impl_fn_ptr_t>{},
177+
td_ns::NullPtrTable<
178+
impl::
179+
binary_contig_row_contig_matrix_broadcast_impl_fn_ptr_t>{});
180+
};
181+
m.def("_fmod", fmod_pyapi, "", py::arg("src1"), py::arg("src2"),
182+
py::arg("dst"), py::arg("sycl_queue"),
183+
py::arg("depends") = py::list());
184+
185+
auto fmod_result_type_pyapi = [&](const py::dtype &dtype1,
186+
const py::dtype &dtype2) {
187+
return py_int::py_binary_ufunc_result_type(
188+
dtype1, dtype2, fmod_output_typeid_table);
189+
};
190+
m.def("_fmod_result_type", fmod_result_type_pyapi);
191+
}
192+
}
193+
} // namespace dpnp::extensions::ufunc

‎dpnp/backend/extensions/ufunc/elementwise_functions/fmod.hpp

Lines changed: 35 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,35 @@
1+
//*****************************************************************************
2+
// Copyright (c) 2024, Intel Corporation
3+
// All rights reserved.
4+
//
5+
// Redistribution and use in source and binary forms, with or without
6+
// modification, are permitted provided that the following conditions are met:
7+
// - Redistributions of source code must retain the above copyright notice,
8+
// this list of conditions and the following disclaimer.
9+
// - Redistributions in binary form must reproduce the above copyright notice,
10+
// this list of conditions and the following disclaimer in the documentation
11+
// and/or other materials provided with the distribution.
12+
//
13+
// THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS"
14+
// AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE
15+
// IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE
16+
// ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE
17+
// LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR
18+
// CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF
19+
// SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS
20+
// INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN
21+
// CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE)
22+
// ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF
23+
// THE POSSIBILITY OF SUCH DAMAGE.
24+
//*****************************************************************************
25+
26+
#pragma once
27+
28+
#include <pybind11/pybind11.h>
29+
30+
namespace py = pybind11;
31+
32+
namespace dpnp::extensions::ufunc
33+
{
34+
void init_fmod(py::module_ m);
35+
} // namespace dpnp::extensions::ufunc

0 commit comments

Comments
 (0)