cuda-api-wrappers
Thin C++-flavored wrappers for the CUDA Runtime API
kernel.hpp
Go to the documentation of this file.
1 
10 #pragma once
11 #ifndef MULTI_WRAPPER_IMPLS_KERNEL_HPP_
12 #define MULTI_WRAPPER_IMPLS_KERNEL_HPP_
13 
14 #include "../device.hpp"
15 #include "../pointer.hpp"
16 #include "../primary_context.hpp"
17 #include "../kernel.hpp"
18 #include "../module.hpp"
19 
20 namespace cuda {
21 
22 inline context_t kernel_t::context() const noexcept
23 {
24  constexpr bool dont_take_ownership { false };
25  return context::wrap(device_id_, context_handle_, dont_take_ownership);
26 }
27 
28 inline device_t kernel_t::device() const noexcept
29 {
30 return device::get(device_id_);
31 }
32 
33 inline void kernel_t::set_attribute(kernel::attribute_t attribute, kernel::attribute_value_t value) const
34 {
35  kernel::detail_::set_attribute_in_current_context(handle_, attribute, value);
36 }
37 
38 #if CUDA_VERSION >= 12030
39 inline module_t kernel_t::module() const
40 {
41  auto module_handle = kernel::detail_::get_module(context_handle_, handle_);
42  return module::detail_::wrap(device_id_, context_handle_, module_handle, do_not_take_ownership);
43 }
44 #endif
45 
46 namespace detail_ {
47 
48 template<typename Kernel>
49 device::primary_context_t get_implicit_primary_context(Kernel)
50 {
51  return device::current::get().primary_context();
52 }
53 
54 template<>
55 inline device::primary_context_t get_implicit_primary_context<kernel_t>(kernel_t kernel)
56 {
57  auto context = kernel.context();
58  auto device = context.device();
59  auto primary_context = device.primary_context();
60  if (context != primary_context) {
61  throw ::std::logic_error("Attempt to launch a kernel associated with a non-primary context without specifying a stream associated with that context.");
62  }
63  return primary_context;
64 }
65 
66 } // namespace detail_
67 
68 } // namespace cuda
69 
70 #endif // MULTI_WRAPPER_IMPLS_KERNEL_HPP_
71 
int attribute_value_t
The uniform type the CUDA driver uses for all kernel attributes; it is typically more appropriate to ...
Definition: types.hpp:988
Wrapper class for a CUDA context.
Definition: context.hpp:244
Definitions and functionality wrapping CUDA APIs.
Definition: array.hpp:22
A class for holding the primary context of a CUDA device.
Definition: primary_context.hpp:112
Wrapper class for a CUDA code module.
Definition: module.hpp:123
device_t device() const noexcept
Get (a proxy for) the device for (a context of) which this kernel is defined.
Definition: kernel.hpp:28
device::primary_context_t primary_context(bool hold_pc_refcount_unit=false) const
Produce a proxy for the device&#39;s primary context - the one used by runtime API calls.
Definition: device.hpp:152
device_t get(id_t id)
Returns a proxy for the CUDA device with a given id.
Definition: device.hpp:837
A non-owning wrapper for CUDA kernels - whether they be __global__ functions compiled apriori...
Definition: kernel.hpp:159
context_t context() const noexcept
Get (a proxy for) the context in which this kernel is defined.
Definition: kernel.hpp:22
Wrapper class for a CUDA device.
Definition: device.hpp:135
CUfunction_attribute attribute_t
Raw CUDA driver selector of a kernel attribute.
Definition: types.hpp:983