Skip to content

Commit 5579edd

Browse files
committed
Added base f/w for host API as an extension
1 parent 380badd commit 5579edd

File tree

7 files changed

+616
-0
lines changed

7 files changed

+616
-0
lines changed

dpnp/CMakeLists.txt

+1
Original file line numberDiff line numberDiff line change
@@ -59,6 +59,7 @@ add_subdirectory(backend)
5959
add_subdirectory(backend/extensions/blas)
6060
add_subdirectory(backend/extensions/lapack)
6161
add_subdirectory(backend/extensions/rng/device)
62+
add_subdirectory(backend/extensions/rng/host)
6263
add_subdirectory(backend/extensions/vm)
6364
add_subdirectory(backend/extensions/sycl_ext)
6465

Original file line numberDiff line numberDiff line change
@@ -0,0 +1,77 @@
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+
27+
set(python_module_name _rng_host_impl)
28+
pybind11_add_module(${python_module_name} MODULE
29+
rng_py.cpp
30+
gaussian.cpp
31+
)
32+
33+
if (WIN32)
34+
if (${CMAKE_VERSION} VERSION_LESS "3.27")
35+
# this is a work-around for target_link_options inserting option after -link option, cause
36+
# linker to ignore it.
37+
set(CMAKE_CXX_LINK_FLAGS "${CMAKE_CXX_LINK_FLAGS} -fsycl-device-code-split=per_kernel")
38+
endif()
39+
endif()
40+
41+
set_target_properties(${python_module_name} PROPERTIES CMAKE_POSITION_INDEPENDENT_CODE ON)
42+
43+
target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR})
44+
target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/engine)
45+
target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../../include)
46+
target_include_directories(${python_module_name} PRIVATE ${CMAKE_CURRENT_SOURCE_DIR}/../../src)
47+
48+
target_include_directories(${python_module_name} PUBLIC ${Dpctl_INCLUDE_DIRS})
49+
target_include_directories(${python_module_name} PUBLIC ${Dpctl_TENSOR_INCLUDE_DIR})
50+
51+
if (WIN32)
52+
target_compile_options(${python_module_name} PRIVATE
53+
/clang:-fno-approx-func
54+
/clang:-fno-finite-math-only
55+
)
56+
else()
57+
target_compile_options(${python_module_name} PRIVATE
58+
-fno-approx-func
59+
-fno-finite-math-only
60+
)
61+
endif()
62+
63+
target_link_options(${python_module_name} PUBLIC -fsycl-device-code-split=per_kernel)
64+
if (UNIX)
65+
# this option is support on Linux only
66+
target_link_options(${python_module_name} PUBLIC -fsycl-link-huge-device-code)
67+
endif()
68+
69+
if (DPNP_GENERATE_COVERAGE)
70+
target_link_options(${python_module_name} PRIVATE -fprofile-instr-generate -fcoverage-mapping)
71+
endif()
72+
73+
target_link_libraries(${python_module_name} PUBLIC MKL::MKL_DPCPP)
74+
75+
install(TARGETS ${python_module_name}
76+
DESTINATION "dpnp/backend/extensions/rng/host"
77+
)
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,65 @@
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 <oneapi/mkl/rng/device.hpp>
29+
30+
#include "utils/type_dispatch.hpp"
31+
32+
namespace dpnp::backend::ext::rng::host::dispatch
33+
{
34+
namespace dpctl_td_ns = dpctl::tensor::type_dispatch;
35+
namespace mkl_rng = oneapi::mkl::rng;
36+
37+
template <typename Ty, typename ArgTy, typename Method, typename argMethod>
38+
struct TypePairDefinedEntry
39+
: std::bool_constant<std::is_same_v<Ty, ArgTy> &&
40+
std::is_same_v<Method, argMethod>>
41+
{
42+
static constexpr bool is_defined = true;
43+
};
44+
45+
template <typename T, typename M>
46+
struct GaussianTypePairSupportFactory
47+
{
48+
static constexpr bool is_defined = std::disjunction<
49+
TypePairDefinedEntry<T,
50+
double,
51+
M,
52+
mkl_rng::gaussian_method::by_default>,
53+
TypePairDefinedEntry<T,
54+
double,
55+
M,
56+
mkl_rng::gaussian_method::box_muller2>,
57+
TypePairDefinedEntry<T, float, M, mkl_rng::gaussian_method::by_default>,
58+
TypePairDefinedEntry<T,
59+
float,
60+
M,
61+
mkl_rng::gaussian_method::box_muller2>,
62+
// fall-through
63+
dpctl_td_ns::NotDefinedEntry>::is_defined;
64+
};
65+
} // namespace dpnp::backend::ext::rng::host::dispatch
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,106 @@
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 <oneapi/mkl/rng.hpp>
29+
30+
namespace dpnp::backend::ext::rng::host::dispatch
31+
{
32+
namespace mkl_rng = oneapi::mkl::rng;
33+
34+
template <typename funcPtrT,
35+
template <typename fnT, typename E, typename T, typename M>
36+
typename factory,
37+
int _no_of_engines,
38+
int _no_of_types,
39+
int _no_of_methods>
40+
class Dispatch3DTableBuilder
41+
{
42+
private:
43+
template <typename E, typename T>
44+
const std::vector<funcPtrT> row_per_method() const
45+
{
46+
std::vector<funcPtrT> per_method = {
47+
factory<funcPtrT, E, T, mkl_rng::gaussian_method::by_default>{}
48+
.get(),
49+
factory<funcPtrT, E, T, mkl_rng::gaussian_method::box_muller2>{}
50+
.get(),
51+
};
52+
assert(per_method.size() == _no_of_methods);
53+
return per_method;
54+
}
55+
56+
template <typename E>
57+
auto table_per_type_and_method() const
58+
{
59+
std::vector<std::vector<funcPtrT>> table_by_type = {
60+
row_per_method<E, bool>(),
61+
row_per_method<E, int8_t>(),
62+
row_per_method<E, uint8_t>(),
63+
row_per_method<E, int16_t>(),
64+
row_per_method<E, uint16_t>(),
65+
row_per_method<E, int32_t>(),
66+
row_per_method<E, uint32_t>(),
67+
row_per_method<E, int64_t>(),
68+
row_per_method<E, uint64_t>(),
69+
row_per_method<E, sycl::half>(),
70+
row_per_method<E, float>(),
71+
row_per_method<E, double>(),
72+
row_per_method<E, std::complex<float>>(),
73+
row_per_method<E, std::complex<double>>()};
74+
assert(table_by_type.size() == _no_of_types);
75+
return table_by_type;
76+
}
77+
78+
public:
79+
Dispatch3DTableBuilder() = default;
80+
~Dispatch3DTableBuilder() = default;
81+
82+
void populate(funcPtrT table[][_no_of_types][_no_of_methods]) const
83+
{
84+
const auto map_by_engine = {
85+
table_per_type_and_method<mkl_rng::mrg32k3a>(),
86+
table_per_type_and_method<mkl_rng::philox4x32x10>(),
87+
table_per_type_and_method<mkl_rng::mcg31m1>(),
88+
table_per_type_and_method<mkl_rng::mcg59>()};
89+
assert(map_by_engine.size() == _no_of_engines);
90+
91+
std::uint16_t engine_id = 0;
92+
for (auto &table_by_type : map_by_engine) {
93+
std::uint16_t type_id = 0;
94+
for (auto &row_by_method : table_by_type) {
95+
std::uint16_t method_id = 0;
96+
for (auto &fn_ptr : row_by_method) {
97+
table[engine_id][type_id][method_id] = fn_ptr;
98+
++method_id;
99+
}
100+
++type_id;
101+
}
102+
++engine_id;
103+
}
104+
}
105+
};
106+
} // namespace dpnp::backend::ext::rng::host::dispatch

0 commit comments

Comments
 (0)