Skip to content

Commit afb84af

Browse files
vtavanavtavana
committed
implement mkl version of add, subtract and multiply
1 parent 47a1ec4 commit afb84af

File tree

8 files changed

+497
-36
lines changed

8 files changed

+497
-36
lines changed

‎dpnp/backend/extensions/vm/add.hpp

+81
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,81 @@
1+
//*****************************************************************************
2+
// Copyright (c) 2023, 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 <CL/sycl.hpp>
29+
30+
#include "common.hpp"
31+
#include "types_matrix.hpp"
32+
33+
namespace dpnp
34+
{
35+
namespace backend
36+
{
37+
namespace ext
38+
{
39+
namespace vm
40+
{
41+
template <typename T>
42+
sycl::event add_contig_impl(sycl::queue exec_q,
43+
const std::int64_t n,
44+
const char *in_a,
45+
const char *in_b,
46+
char *out_y,
47+
const std::vector<sycl::event> &depends)
48+
{
49+
type_utils::validate_type_for_device<T>(exec_q);
50+
51+
const T *a = reinterpret_cast<const T *>(in_a);
52+
const T *b = reinterpret_cast<const T *>(in_b);
53+
T *y = reinterpret_cast<T *>(out_y);
54+
55+
return mkl_vm::add(exec_q,
56+
n, // number of elements to be calculated
57+
a, // pointer `a` containing 1st input vector of size n
58+
b, // pointer `b` containing 2nd input vector of size n
59+
y, // pointer `y` to the output vector of size n
60+
depends);
61+
}
62+
63+
template <typename fnT, typename T>
64+
struct AddContigFactory
65+
{
66+
fnT get()
67+
{
68+
if constexpr (std::is_same_v<
69+
typename types::AddOutputType<T>::value_type, void>)
70+
{
71+
return nullptr;
72+
}
73+
else {
74+
return add_contig_impl<T>;
75+
}
76+
}
77+
};
78+
} // namespace vm
79+
} // namespace ext
80+
} // namespace backend
81+
} // namespace dpnp

‎dpnp/backend/extensions/vm/mul.hpp

+81
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,81 @@
1+
//*****************************************************************************
2+
// Copyright (c) 2023, 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 <CL/sycl.hpp>
29+
30+
#include "common.hpp"
31+
#include "types_matrix.hpp"
32+
33+
namespace dpnp
34+
{
35+
namespace backend
36+
{
37+
namespace ext
38+
{
39+
namespace vm
40+
{
41+
template <typename T>
42+
sycl::event mul_contig_impl(sycl::queue exec_q,
43+
const std::int64_t n,
44+
const char *in_a,
45+
const char *in_b,
46+
char *out_y,
47+
const std::vector<sycl::event> &depends)
48+
{
49+
type_utils::validate_type_for_device<T>(exec_q);
50+
51+
const T *a = reinterpret_cast<const T *>(in_a);
52+
const T *b = reinterpret_cast<const T *>(in_b);
53+
T *y = reinterpret_cast<T *>(out_y);
54+
55+
return mkl_vm::mul(exec_q,
56+
n, // number of elements to be calculated
57+
a, // pointer `a` containing 1st input vector of size n
58+
b, // pointer `b` containing 2nd input vector of size n
59+
y, // pointer `y` to the output vector of size n
60+
depends);
61+
}
62+
63+
template <typename fnT, typename T>
64+
struct MulContigFactory
65+
{
66+
fnT get()
67+
{
68+
if constexpr (std::is_same_v<
69+
typename types::MulOutputType<T>::value_type, void>)
70+
{
71+
return nullptr;
72+
}
73+
else {
74+
return mul_contig_impl<T>;
75+
}
76+
}
77+
};
78+
} // namespace vm
79+
} // namespace ext
80+
} // namespace backend
81+
} // namespace dpnp

‎dpnp/backend/extensions/vm/sub.hpp

+81
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,81 @@
1+
//*****************************************************************************
2+
// Copyright (c) 2023, 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 <CL/sycl.hpp>
29+
30+
#include "common.hpp"
31+
#include "types_matrix.hpp"
32+
33+
namespace dpnp
34+
{
35+
namespace backend
36+
{
37+
namespace ext
38+
{
39+
namespace vm
40+
{
41+
template <typename T>
42+
sycl::event sub_contig_impl(sycl::queue exec_q,
43+
const std::int64_t n,
44+
const char *in_a,
45+
const char *in_b,
46+
char *out_y,
47+
const std::vector<sycl::event> &depends)
48+
{
49+
type_utils::validate_type_for_device<T>(exec_q);
50+
51+
const T *a = reinterpret_cast<const T *>(in_a);
52+
const T *b = reinterpret_cast<const T *>(in_b);
53+
T *y = reinterpret_cast<T *>(out_y);
54+
55+
return mkl_vm::sub(exec_q,
56+
n, // number of elements to be calculated
57+
a, // pointer `a` containing 1st input vector of size n
58+
b, // pointer `b` containing 2nd input vector of size n
59+
y, // pointer `y` to the output vector of size n
60+
depends);
61+
}
62+
63+
template <typename fnT, typename T>
64+
struct SubContigFactory
65+
{
66+
fnT get()
67+
{
68+
if constexpr (std::is_same_v<
69+
typename types::SubOutputType<T>::value_type, void>)
70+
{
71+
return nullptr;
72+
}
73+
else {
74+
return sub_contig_impl<T>;
75+
}
76+
}
77+
};
78+
} // namespace vm
79+
} // namespace ext
80+
} // namespace backend
81+
} // namespace dpnp

‎dpnp/backend/extensions/vm/types_matrix.hpp

+75
Original file line numberDiff line numberDiff line change
@@ -43,6 +43,31 @@ namespace vm
4343
{
4444
namespace types
4545
{
46+
/**
47+
* @brief A factory to define pairs of supported types for which
48+
* MKL VM library provides support in oneapi::mkl::vm::add<T> function.
49+
*
50+
* @tparam T Type of input vectors `a` and `b` and of result vector `y`.
51+
*/
52+
template <typename T>
53+
struct AddOutputType
54+
{
55+
using value_type = typename std::disjunction<
56+
dpctl_td_ns::BinaryTypeMapResultEntry<T,
57+
std::complex<double>,
58+
T,
59+
std::complex<double>,
60+
std::complex<double>>,
61+
dpctl_td_ns::BinaryTypeMapResultEntry<T,
62+
std::complex<float>,
63+
T,
64+
std::complex<float>,
65+
std::complex<float>>,
66+
dpctl_td_ns::BinaryTypeMapResultEntry<T, double, T, double, double>,
67+
dpctl_td_ns::BinaryTypeMapResultEntry<T, float, T, float, float>,
68+
dpctl_td_ns::DefaultResultEntry<void>>::result_type;
69+
};
70+
4671
/**
4772
* @brief A factory to define pairs of supported types for which
4873
* MKL VM library provides support in oneapi::mkl::vm::div<T> function.
@@ -83,6 +108,56 @@ struct CeilOutputType
83108
dpctl_td_ns::DefaultResultEntry<void>>::result_type;
84109
};
85110

111+
/**
112+
* @brief A factory to define pairs of supported types for which
113+
* MKL VM library provides support in oneapi::mkl::vm::mul<T> function.
114+
*
115+
* @tparam T Type of input vectors `a` and `b` and of result vector `y`.
116+
*/
117+
template <typename T>
118+
struct MulOutputType
119+
{
120+
using value_type = typename std::disjunction<
121+
dpctl_td_ns::BinaryTypeMapResultEntry<T,
122+
std::complex<double>,
123+
T,
124+
std::complex<double>,
125+
std::complex<double>>,
126+
dpctl_td_ns::BinaryTypeMapResultEntry<T,
127+
std::complex<float>,
128+
T,
129+
std::complex<float>,
130+
std::complex<float>>,
131+
dpctl_td_ns::BinaryTypeMapResultEntry<T, double, T, double, double>,
132+
dpctl_td_ns::BinaryTypeMapResultEntry<T, float, T, float, float>,
133+
dpctl_td_ns::DefaultResultEntry<void>>::result_type;
134+
};
135+
136+
/**
137+
* @brief A factory to define pairs of supported types for which
138+
* MKL VM library provides support in oneapi::mkl::vm::sub<T> function.
139+
*
140+
* @tparam T Type of input vectors `a` and `b` and of result vector `y`.
141+
*/
142+
template <typename T>
143+
struct SubOutputType
144+
{
145+
using value_type = typename std::disjunction<
146+
dpctl_td_ns::BinaryTypeMapResultEntry<T,
147+
std::complex<double>,
148+
T,
149+
std::complex<double>,
150+
std::complex<double>>,
151+
dpctl_td_ns::BinaryTypeMapResultEntry<T,
152+
std::complex<float>,
153+
T,
154+
std::complex<float>,
155+
std::complex<float>>,
156+
dpctl_td_ns::BinaryTypeMapResultEntry<T, double, T, double, double>,
157+
dpctl_td_ns::BinaryTypeMapResultEntry<T, float, T, float, float>,
158+
dpctl_td_ns::DefaultResultEntry<void>>::result_type;
159+
};
160+
86161
/**
87162
* @brief A factory to define pairs of supported types for which
88163
* MKL VM library provides support in oneapi::mkl::vm::cos<T> function.

0 commit comments

Comments
 (0)