DPC++ Runtime
Runtime libraries for oneAPI DPC++
kernel_impl.cpp
Go to the documentation of this file.
1
//==------- kernel_impl.cpp --- SYCL kernel implementation -----------------==//
2
//
3
// Part of the LLVM Project, under the Apache License v2.0 with LLVM Exceptions.
4
// See https://llvm.org/LICENSE.txt for license information.
5
// SPDX-License-Identifier: Apache-2.0 WITH LLVM-exception
6
//
7
//===----------------------------------------------------------------------===//
8
9
#include <
detail/context_impl.hpp
>
10
#include <
detail/kernel_bundle_impl.hpp
>
11
#include <
detail/kernel_impl.hpp
>
12
#include <
detail/program_impl.hpp
>
13
14
#include <memory>
15
16
__SYCL_INLINE_NAMESPACE
(
cl
) {
17
namespace
sycl
{
18
namespace
detail {
19
20
kernel_impl::kernel_impl(
RT::PiKernel
Kernel,
ContextImplPtr
Context,
21
KernelBundleImplPtr
KernelBundleImpl)
22
:
kernel_impl
(Kernel, Context,
23
std
::make_shared<
program_impl
>(Context, Kernel),
24
/*IsCreatedFromSource*/
true, KernelBundleImpl) {
25
// This constructor is only called in the interoperability kernel constructor.
26
// Let the runtime caller handle native kernel retaining in other cases if
27
// it's needed.
28
getPlugin
().
call
<
PiApiKind::piKernelRetain
>(MKernel);
29
// Enable USM indirect access for interoperability kernels.
30
// Some PI Plugins (like OpenCL) require this call to enable USM
31
// For others, PI will turn this into a NOP.
32
getPlugin
().
call
<
PiApiKind::piKernelSetExecInfo
>(
33
MKernel,
PI_USM_INDIRECT_ACCESS
,
sizeof
(
pi_bool
), &
PI_TRUE
);
34
35
MIsInterop =
true
;
36
}
37
38
kernel_impl::kernel_impl
(
RT::PiKernel
Kernel,
ContextImplPtr
ContextImpl,
39
ProgramImplPtr
ProgramImpl,
bool
IsCreatedFromSource,
40
KernelBundleImplPtr
KernelBundleImpl)
41
: MKernel(Kernel), MContext(ContextImpl),
42
MProgramImpl(
std
::move(ProgramImpl)),
43
MCreatedFromSource(IsCreatedFromSource),
44
MKernelBundleImpl(
std
::move(KernelBundleImpl)) {
45
46
RT::PiContext
Context =
nullptr
;
47
// Using the plugin from the passed ContextImpl
48
getPlugin
().
call
<
PiApiKind::piKernelGetInfo
>(
49
MKernel,
PI_KERNEL_INFO_CONTEXT
,
sizeof
(Context), &Context,
nullptr
);
50
if
(ContextImpl->getHandleRef() != Context)
51
throw
cl::sycl::invalid_parameter_error(
52
"Input context must be the same as the context of cl_kernel"
,
53
PI_INVALID_CONTEXT
);
54
55
MIsInterop = MProgramImpl->isInterop();
56
}
57
58
kernel_impl::kernel_impl
(
RT::PiKernel
Kernel,
ContextImplPtr
ContextImpl,
59
DeviceImageImplPtr
DeviceImageImpl,
60
KernelBundleImplPtr
KernelBundleImpl)
61
: MKernel(Kernel), MContext(
std
::move(ContextImpl)), MProgramImpl(nullptr),
62
MCreatedFromSource(false), MDeviceImageImpl(
std
::move(DeviceImageImpl)),
63
MKernelBundleImpl(
std
::move(KernelBundleImpl)) {
64
65
// kernel_impl shared ownership of kernel handle
66
if
(!
is_host
()) {
67
getPlugin
().
call
<
PiApiKind::piKernelRetain
>(MKernel);
68
}
69
70
MIsInterop = MKernelBundleImpl->isInterop();
71
}
72
73
kernel_impl::kernel_impl
(
ContextImplPtr
Context,
ProgramImplPtr
ProgramImpl)
74
: MContext(Context), MProgramImpl(
std
::move(ProgramImpl)) {}
75
76
kernel_impl::~kernel_impl
() {
77
// TODO catch an exception and put it to list of asynchronous exceptions
78
if
(!
is_host
()) {
79
getPlugin
().
call
<
PiApiKind::piKernelRelease
>(MKernel);
80
}
81
}
82
83
84
bool
kernel_impl::isCreatedFromSource
()
const
{
85
// TODO it is not clear how to understand whether the SYCL kernel is created
86
// from source code or not when the SYCL kernel is created using
87
// the interoperability constructor.
88
// Here a strange case which does not work now:
89
// context Context;
90
// program Program(Context);
91
// Program.build_with_kernel_type<class A>();
92
// kernel FirstKernel= Program.get_kernel<class A>();
93
// cl_kernel ClKernel = FirstKernel.get();
94
// kernel SecondKernel = kernel(ClKernel, Context);
95
// clReleaseKernel(ClKernel);
96
// FirstKernel.isCreatedFromSource() != FirstKernel.isCreatedFromSource();
97
return
MCreatedFromSource;
98
}
99
100
}
// namespace detail
101
}
// namespace sycl
102
}
// __SYCL_INLINE_NAMESPACE(cl)
cl::sycl::detail::ContextImplPtr
std::shared_ptr< detail::context_impl > ContextImplPtr
Definition:
memory_manager.hpp:32
pi_bool
pi_uint32 pi_bool
Definition:
pi.h:74
context_impl.hpp
cl::sycl::detail::kernel_impl::isCreatedFromSource
bool isCreatedFromSource() const
Check if kernel was created from a program that had been created from source.
Definition:
kernel_impl.cpp:84
cl::sycl::detail::kernel_impl::kernel_impl
kernel_impl(RT::PiKernel Kernel, ContextImplPtr Context, KernelBundleImplPtr KernelBundleImpl)
Constructs a SYCL kernel instance from a PiKernel.
Definition:
kernel_impl.cpp:20
cl::sycl::detail::program_impl
Definition:
program_impl.hpp:37
cl::sycl::detail::KernelBundleImplPtr
std::shared_ptr< detail::kernel_bundle_impl > KernelBundleImplPtr
Definition:
kernel_bundle.hpp:132
sycl
Definition:
invoke_simd.hpp:68
PI_USM_INDIRECT_ACCESS
@ PI_USM_INDIRECT_ACCESS
indicates that the kernel might access data through USM ptrs
Definition:
pi.h:1327
cl::sycl::detail::kernel_impl
Definition:
kernel_impl.hpp:34
piKernelRelease
pi_result piKernelRelease(pi_kernel kernel)
Definition:
pi_esimd_emulator.cpp:1383
_pi_kernel
Implementation of a PI Kernel for CUDA.
Definition:
pi_cuda.hpp:624
program_impl.hpp
PI_KERNEL_INFO_CONTEXT
@ PI_KERNEL_INFO_CONTEXT
Definition:
pi.h:363
piKernelSetExecInfo
pi_result piKernelSetExecInfo(pi_kernel kernel, pi_kernel_exec_info value_name, size_t param_value_size, const void *param_value)
API to set attributes controlling kernel execution.
Definition:
pi_esimd_emulator.cpp:1937
cl::sycl::detail::plugin::call
void call(ArgsT... Args) const
Calls the API, traces the call, checks the result.
Definition:
plugin.hpp:217
piKernelRetain
pi_result piKernelRetain(pi_kernel kernel)
Definition:
pi_esimd_emulator.cpp:1381
cl
We provide new interfaces for matrix muliply in this patch:
Definition:
access.hpp:13
PI_INVALID_CONTEXT
@ PI_INVALID_CONTEXT
Definition:
pi.h:92
piKernelGetInfo
pi_result piKernelGetInfo(pi_kernel kernel, pi_kernel_info param_name, size_t param_value_size, void *param_value, size_t *param_value_size_ret)
Definition:
pi_esimd_emulator.cpp:1366
cl::sycl::detail::kernel_impl::is_host
bool is_host() const
Check if the associated SYCL context is a SYCL host context.
Definition:
kernel_impl.hpp:112
std
Definition:
accessor.hpp:2616
cl::sycl::detail::kernel_impl::~kernel_impl
~kernel_impl()
Definition:
kernel_impl.cpp:76
kernel_bundle_impl.hpp
kernel_impl.hpp
PI_TRUE
const pi_bool PI_TRUE
Definition:
pi.h:514
cl::sycl::detail::ProgramImplPtr
std::shared_ptr< program_impl > ProgramImplPtr
Definition:
kernel_impl.hpp:32
_pi_context
PI context mapping to a CUDA context object.
Definition:
pi_cuda.hpp:150
__SYCL_INLINE_NAMESPACE
#define __SYCL_INLINE_NAMESPACE(X)
Definition:
defines_elementary.hpp:12
cl::sycl::detail::DeviceImageImplPtr
std::shared_ptr< device_image_impl > DeviceImageImplPtr
Definition:
kernel_bundle.hpp:67
cl::sycl::detail::kernel_impl::getPlugin
const plugin & getPlugin() const
Definition:
kernel_impl.hpp:114
source
detail
kernel_impl.cpp
Generated by
1.8.17