kernel_launch.hpp File Reference

Variadic, chevron-less wrappers for the CUDA kernel launch mechanism. More...

#include "launch_configuration.hpp"
#include "kernel.hpp"
#include "kernels/apriori_compiled.hpp"
#include <type_traits>
#include <utility>

Include dependency graph for kernel_launch.hpp:

This graph shows which files directly or indirectly include this file:

Go to the source code of this file.

Namespaces
	cuda
	Definitions and functionality wrapping CUDA APIs.

Functions
constexpr grid::dimensions_t	cuda::single_block ()
	A named constructor idiom for grid::dimensions_t, which, when used, will result in a grid with a single block.
constexpr grid::block_dimensions_t	cuda::single_thread_per_block ()
	A named constructor idiom for grid::dimensions_t, which, when used, will result in a grid whose blocks have a single thread.
template<typename... KernelParameters>
detail_::raw_kernel_typegen< KernelParameters... >::type	cuda::kernel::unwrap (const kernel::apriori_compiled_t &kernel)
	A function similar to ::std::any_cast for retrieving the function pointer wrapped by a cuda::kernel::apriori_compiled_t object: Only the user knows the exact set of kernel function parameters, and the must supply them as template arguments to obtain the function pointer they are after.
template<typename Kernel , typename... KernelParameters>
void	cuda::enqueue_launch (Kernel &&kernel, const stream_t &stream, launch_configuration_t launch_configuration, KernelParameters &&... parameters)
	Enqueues a kernel on a stream (=queue) on the current CUDA device. More...
template<typename Kernel , typename... KernelParameters>
void	cuda::launch (Kernel &&kernel, launch_configuration_t launch_configuration, KernelParameters &&... parameters)
	Variant of enqueue_launch for use with the default stream in the current context. More...
template<typename SpanOfConstVoidPtrLike >
void	cuda::launch_type_erased (const kernel_t &kernel, const stream_t &stream, launch_configuration_t launch_configuration, SpanOfConstVoidPtrLike marshalled_arguments)
	Launch a kernel with the arguments pre-marshalled into the (main) form which the CUDA driver's launch primitive accepts variables in: A null- terminated sequence of (possibly const) void *'s to the argument values. More...

Detailed Description

Variadic, chevron-less wrappers for the CUDA kernel launch mechanism.

This file has two stand-alone functions used for launching kernels - by application code directly and by other API wrappers (e.g. cuda::device_t and cuda::stream_t ).

The wrapper functions have two goals:

• Avoiding the annoying triple-chevron syntax, e.g.

  my_kernel<<<launch, config, stuff>>>(real, args)

  and sticking to proper C++; in other words, the wrappers are "ugly" instead of client code having to be.

• Avoiding some of the "parameter soup" of launching a kernel: It's not so difficult to mix up shared memory sizes with stream handles; grid and block dimensions with each other; and even grid/block dimensions with the scalar parameters - since a dim3 is constructible from integral values. Instead, we enforce a launch configuration structure: cuda::launch_configuration_t .

Note

Even though when you use this wrapper, your code will not have the silly chevron, you can't use it from regular .cpp files compiled with your host compiler. Hence the .cuh extension. You can, however, safely include this file from your .cpp for other definitions.