17 #ifndef KOKKOS_IMPL_PUBLIC_INCLUDE
18 #include <Kokkos_Macros.hpp>
20 "Including non-public Kokkos header files is not allowed.");
22 #ifndef KOKKOS_SYCL_HPP
23 #define KOKKOS_SYCL_HPP
25 #include <Kokkos_Macros.hpp>
27 #ifdef KOKKOS_ENABLE_SYCL
29 #if __has_include(<sycl/sycl.hpp>)
30 #include <sycl/sycl.hpp>
32 #include <CL/sycl.hpp>
34 #include <Kokkos_SYCL_Space.hpp>
35 #include <Kokkos_Layout.hpp>
36 #include <Kokkos_ScratchSpace.hpp>
37 #include <impl/Kokkos_Profiling_Interface.hpp>
38 #include <impl/Kokkos_HostSharedPtr.hpp>
39 #include <impl/Kokkos_InitializationSettings.hpp>
42 namespace Experimental {
56 using execution_space = SYCL;
57 using memory_space = SYCLDeviceUSMSpace;
58 using device_type = Kokkos::Device<execution_space, memory_space>;
60 using array_layout = LayoutLeft;
61 using size_type = memory_space::size_type;
63 using scratch_memory_space = ScratchMemorySpace<SYCL>;
66 explicit SYCL(
const sycl::queue&);
68 uint32_t impl_instance_id() const noexcept {
69 return m_space_instance->impl_get_instance_id();
72 sycl::queue& sycl_queue() const noexcept {
73 return *m_space_instance->m_queue;
81 KOKKOS_INLINE_FUNCTION
static int in_parallel() {
82 #if defined(__SYCL_DEVICE_ONLY__)
96 static void impl_static_fence(
const std::string& name);
99 const std::string& name =
100 "Kokkos::Experimental::SYCL::fence: Unnamed Instance Fence")
const;
103 void print_configuration(std::ostream& os,
bool verbose =
false)
const;
106 static void impl_finalize();
108 static void impl_initialize(InitializationSettings
const&);
110 static bool impl_is_initialized();
112 #ifdef KOKKOS_ENABLE_DEPRECATED_CODE_4
113 static int concurrency();
115 int concurrency()
const;
118 static const char* name();
120 inline Impl::SYCLInternal* impl_internal_space_instance()
const {
121 return m_space_instance.get();
125 static std::ostream& impl_sycl_info(std::ostream& os,
126 const sycl::device& device);
128 friend bool operator==(SYCL
const& lhs, SYCL
const& rhs) {
129 return lhs.impl_internal_space_instance() ==
130 rhs.impl_internal_space_instance();
132 friend bool operator!=(SYCL
const& lhs, SYCL
const& rhs) {
133 return !(lhs == rhs);
135 Kokkos::Impl::HostSharedPtr<Impl::SYCLInternal> m_space_instance;
141 namespace Experimental {
143 struct DeviceTypeTraits<Kokkos::Experimental::SYCL> {
145 static constexpr DeviceType
id = DeviceType::SYCL;
146 static int device_id(
const Kokkos::Experimental::SYCL& exec) {
147 return exec.impl_internal_space_instance()->m_syclDev;
153 namespace Experimental {
154 template <
class... Args>
155 std::vector<SYCL> partition_space(
const SYCL& sycl_space, Args...) {
157 (... && std::is_arithmetic_v<Args>),
158 "Kokkos Error: partitioning arguments must be integers or floats");
160 sycl::context context = sycl_space.sycl_queue().get_context();
161 sycl::device device =
162 sycl_space.impl_internal_space_instance()->m_queue->get_device();
163 std::vector<SYCL> instances;
164 instances.reserve(
sizeof...(Args));
165 for (
unsigned int i = 0; i <
sizeof...(Args); ++i)
166 instances.emplace_back(sycl::queue(context, device));
171 std::vector<SYCL> partition_space(
const SYCL& sycl_space,
172 std::vector<T>& weights) {
174 std::is_arithmetic<T>::value,
175 "Kokkos Error: partitioning arguments must be integers or floats");
177 sycl::context context = sycl_space.sycl_queue().get_context();
178 sycl::device device =
179 sycl_space.impl_internal_space_instance()->m_queue->get_device();
180 std::vector<SYCL> instances;
181 instances.reserve(weights.size());
182 for (
unsigned int i = 0; i < weights.size(); ++i)
183 instances.emplace_back(sycl::queue(context, device));