/// Enqueues a command to fill \p image with \p fill_color.
///
/// \see_opencl_ref{clEnqueueFillImage}
///
/// \opencl_version_warning{1,2}
event enqueue_fill_image(const image3d &image,
const void *fill_color,
const size_t origin[3],
const size_t region[3],
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(image.get_context() == this->get_context());
event event_;
cl_int ret = clEnqueueFillImage(
m_queue,
image.get(),
fill_color,
origin,
region,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to read data from \p buffer to host memory.
///
/// \see_opencl_ref{clEnqueueReadBuffer}
///
/// \see copy()
void enqueue_read_buffer(const buffer &buffer,
size_t offset,
size_t size,
void *host_ptr,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(size <= buffer.size());
BOOST_ASSERT(buffer.get_context() == this->get_context());
BOOST_ASSERT(host_ptr != 0);
cl_int ret = clEnqueueReadBuffer(
m_queue,
buffer.get(),
CL_TRUE,
offset,
size,
host_ptr,
events.size(),
events.get_event_ptr(),
0
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
}
/// Enqueues a command to write data from host memory to \p buffer.
/// The copy is performed asynchronously.
///
/// \see_opencl_ref{clEnqueueWriteBuffer}
///
/// \see copy_async()
event enqueue_write_buffer_async(const buffer &buffer,
size_t offset,
size_t size,
const void *host_ptr,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(size <= buffer.size());
BOOST_ASSERT(buffer.get_context() == this->get_context());
BOOST_ASSERT(host_ptr != 0);
event event_;
cl_int ret = clEnqueueWriteBuffer(
m_queue,
buffer.get(),
CL_FALSE,
offset,
size,
host_ptr,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to read data from \p image to host memory.
///
/// \see_opencl_ref{clEnqueueReadImage}
void enqueue_read_image(const image2d &image,
const size_t origin[2],
const size_t region[2],
size_t row_pitch,
void *host_ptr,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(image.get_context() == this->get_context());
const size_t origin3[3] = { origin[0], origin[1], size_t(0) };
const size_t region3[3] = { region[0], region[1], size_t(1) };
cl_int ret = clEnqueueReadImage(
m_queue,
image.get(),
CL_TRUE,
origin3,
region3,
row_pitch,
0,
host_ptr,
events.size(),
events.get_event_ptr(),
0
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
}
/// Enqueues a command to fill \p buffer with \p pattern.
///
/// \see_opencl_ref{clEnqueueFillBuffer}
///
/// \opencl_version_warning{1,2}
///
/// \see fill()
event enqueue_fill_buffer(const buffer &buffer,
const void *pattern,
size_t pattern_size,
size_t offset,
size_t size,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(offset + size <= buffer.size());
BOOST_ASSERT(buffer.get_context() == this->get_context());
event event_;
cl_int ret = clEnqueueFillBuffer(
m_queue,
buffer.get(),
pattern,
pattern_size,
offset,
size,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to write data from host memory to \p image.
///
/// \see_opencl_ref{clEnqueueWriteImage}
void enqueue_write_image(const image3d &image,
const size_t origin[3],
const size_t region[3],
size_t input_row_pitch,
size_t input_slice_pitch,
const void *host_ptr,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(image.get_context() == this->get_context());
cl_int ret = clEnqueueWriteImage(
m_queue,
image.get(),
CL_TRUE,
origin,
region,
input_row_pitch,
input_slice_pitch,
host_ptr,
events.size(),
events.get_event_ptr(),
0
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
}
/// Enqueues a command to fill \p size bytes of data at \p svm_ptr with
/// \p pattern.
///
/// \opencl_version_warning{2,0}
///
/// \see_opencl2_ref{clEnqueueSVMMemFill}
event enqueue_svm_fill(void *svm_ptr,
const void *pattern,
size_t pattern_size,
size_t size,
const wait_list &events = wait_list())
{
event event_;
cl_int ret = clEnqueueSVMMemFill(
m_queue,
svm_ptr,
pattern,
pattern_size,
size,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to copy data from \p src_image to \p dst_buffer.
///
/// \see_opencl_ref{clEnqueueCopyImageToBuffer}
event enqueue_copy_image_to_buffer(const image2d &src_image,
const buffer &dst_buffer,
const size_t src_origin[2],
const size_t region[2],
size_t dst_offset,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(src_image.get_context() == this->get_context());
BOOST_ASSERT(dst_buffer.get_context() == this->get_context());
const size_t src_origin3[3] = { src_origin[0], src_origin[1], size_t(0) };
const size_t region3[3] = { region[0], region[1], size_t(1) };
event event_;
cl_int ret = clEnqueueCopyImageToBuffer(
m_queue,
src_image.get(),
dst_buffer.get(),
src_origin3,
region3,
dst_offset,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to copy data from \p src_image to \p dst_image.
///
/// \see_opencl_ref{clEnqueueCopyImage}
event enqueue_copy_image(const image3d &src_image,
const image3d &dst_image,
const size_t src_origin[3],
const size_t dst_origin[3],
const size_t region[3],
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(src_image.get_context() == this->get_context());
BOOST_ASSERT(dst_image.get_context() == this->get_context());
BOOST_ASSERT_MSG(src_image.get_format() == dst_image.get_format(),
"Source and destination image formats must match.");
event event_;
cl_int ret = clEnqueueCopyImage(
m_queue,
src_image.get(),
dst_image.get(),
src_origin,
dst_origin,
region,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a function to execute on the host.
event enqueue_native_kernel(void (BOOST_COMPUTE_CL_CALLBACK *user_func)(void *),
void *args,
size_t cb_args,
uint_ num_mem_objects,
const cl_mem *mem_list,
const void **args_mem_loc,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
event event_;
cl_int ret = clEnqueueNativeKernel(
m_queue,
user_func,
args,
cb_args,
num_mem_objects,
mem_list,
args_mem_loc,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a kernel to execute using a single work-item.
///
/// \see_opencl_ref{clEnqueueTask}
event enqueue_task(const kernel &kernel, const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(kernel.get_context() == this->get_context());
event event_;
// clEnqueueTask() was deprecated in OpenCL 2.0. In that case we
// just forward to the equivalent clEnqueueNDRangeKernel() call.
#ifdef CL_VERSION_2_0
size_t one = 1;
cl_int ret = clEnqueueNDRangeKernel(
m_queue, kernel, 1, 0, &one, &one,
events.size(), events.get_event_ptr(), &event_.get()
);
#else
cl_int ret = clEnqueueTask(
m_queue, kernel, events.size(), events.get_event_ptr(), &event_.get()
);
#endif
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a kernel for execution.
///
/// \see_opencl_ref{clEnqueueNDRangeKernel}
event enqueue_nd_range_kernel(const kernel &kernel,
size_t work_dim,
const size_t *global_work_offset,
const size_t *global_work_size,
const size_t *local_work_size,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(kernel.get_context() == this->get_context());
event event_;
cl_int ret = clEnqueueNDRangeKernel(
m_queue,
kernel,
static_cast<cl_uint>(work_dim),
global_work_offset,
global_work_size,
local_work_size,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to migrate \p mem_objects.
///
/// \see_opencl_ref{clEnqueueMigrateMemObjects}
///
/// \opencl_version_warning{1,2}
event enqueue_migrate_memory_objects(uint_ num_mem_objects,
const cl_mem *mem_objects,
cl_mem_migration_flags flags,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
event event_;
cl_int ret = clEnqueueMigrateMemObjects(
m_queue,
num_mem_objects,
mem_objects,
flags,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to copy data from \p src_buffer to
/// \p dst_buffer.
///
/// \see_opencl_ref{clEnqueueCopyBuffer}
///
/// \see copy()
event enqueue_copy_buffer(const buffer &src_buffer,
const buffer &dst_buffer,
size_t src_offset,
size_t dst_offset,
size_t size,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(src_offset + size <= src_buffer.size());
BOOST_ASSERT(dst_offset + size <= dst_buffer.size());
BOOST_ASSERT(src_buffer.get_context() == this->get_context());
BOOST_ASSERT(dst_buffer.get_context() == this->get_context());
event event_;
cl_int ret = clEnqueueCopyBuffer(
m_queue,
src_buffer.get(),
dst_buffer.get(),
src_offset,
dst_offset,
size,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to copy data from \p src_buffer to \p dst_image.
///
/// \see_opencl_ref{clEnqueueCopyBufferToImage}
event enqueue_copy_buffer_to_image(const buffer &src_buffer,
const image3d &dst_image,
size_t src_offset,
const size_t dst_origin[3],
const size_t region[3],
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(src_buffer.get_context() == this->get_context());
BOOST_ASSERT(dst_image.get_context() == this->get_context());
event event_;
cl_int ret = clEnqueueCopyBufferToImage(
m_queue,
src_buffer.get(),
dst_image.get(),
src_offset,
dst_origin,
region,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a barrier in the queue after \p events.
///
/// \opencl_version_warning{1,2}
void enqueue_barrier(const wait_list &events)
{
BOOST_ASSERT(m_queue != 0);
clEnqueueBarrierWithWaitList(
m_queue, events.size(), events.get_event_ptr(), 0
);
}
/// Wait for events in the list
inline void wait_for_events(const wait_list &events) {
std::map<context_id, wait_list> by_context;
for(auto e = events.begin(); e != events.end(); ++e) {
context_id ctx = get_context_id(*e);
by_context[ctx].push_back(*e);
}
for(auto c = by_context.begin(); c != by_context.end(); ++c)
cl::Event::waitForEvents(c->second);
}
/// Enqueue marker (with wait list) into the queue
inline event enqueue_marker(command_queue &q,
const wait_list &events = wait_list())
{
context_id ctx = get_context_id(q);
wait_list my_events;
std::copy_if(events.begin(), events.end(),
std::back_inserter(my_events),
[=](const event &e){ return ctx == get_context_id(e); });
event e;
q.enqueueMarkerWithWaitList(&my_events, &e);
return e;
}
/// Enqueues a marker after \p events in the queue and returns an
/// event that can be used to track its progress.
///
/// \opencl_version_warning{1,2}
event enqueue_marker(const wait_list &events)
{
event event_;
cl_int ret = clEnqueueMarkerWithWaitList(
m_queue, events.size(), events.get_event_ptr(), &event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to unmap \p svm_ptr from the host memory space.
///
/// \opencl_version_warning{2,0}
///
/// \see_opencl2_ref{clEnqueueSVMUnmap}
event enqueue_svm_unmap(void *svm_ptr,
const wait_list &events = wait_list())
{
event event_;
cl_int ret = clEnqueueSVMUnmap(
m_queue,
svm_ptr,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to copy \p size bytes of data from \p src_ptr to
/// \p dst_ptr.
///
/// \opencl_version_warning{2,0}
///
/// \see_opencl2_ref{clEnqueueSVMMemcpy}
void enqueue_svm_memcpy(void *dst_ptr,
const void *src_ptr,
size_t size,
const wait_list &events = wait_list())
{
cl_int ret = clEnqueueSVMMemcpy(
m_queue,
CL_TRUE,
dst_ptr,
src_ptr,
size,
events.size(),
events.get_event_ptr(),
0
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
}
/// Enqueues a command to map \p svm_ptr to the host memory space.
///
/// \opencl_version_warning{2,0}
///
/// \see_opencl2_ref{clEnqueueSVMMap}
void enqueue_svm_map(void *svm_ptr,
size_t size,
cl_map_flags flags,
const wait_list &events = wait_list())
{
cl_int ret = clEnqueueSVMMap(
m_queue,
CL_TRUE,
flags,
svm_ptr,
size,
events.size(),
events.get_event_ptr(),
0
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
}
/// Enqueues a command to release the specified OpenGL memory objects.
///
/// \see_opencl_ref{clEnqueueReleaseGLObjects}
inline event opengl_enqueue_release_gl_objects(size_t num_objects,
const cl_mem *mem_objects,
command_queue &queue,
const wait_list &events = wait_list())
{
BOOST_ASSERT(queue != 0);
event event_;
cl_int ret = clEnqueueReleaseGLObjects(queue.get(),
num_objects,
mem_objects,
events.size(),
events.get_event_ptr(),
&event_.get());
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to copy a rectangular region from
/// \p src_buffer to \p dst_buffer.
///
/// \see_opencl_ref{clEnqueueCopyBufferRect}
///
/// \opencl_version_warning{1,1}
event enqueue_copy_buffer_rect(const buffer &src_buffer,
const buffer &dst_buffer,
const size_t src_origin[3],
const size_t dst_origin[3],
const size_t region[3],
size_t buffer_row_pitch,
size_t buffer_slice_pitch,
size_t host_row_pitch,
size_t host_slice_pitch,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(src_buffer.get_context() == this->get_context());
BOOST_ASSERT(dst_buffer.get_context() == this->get_context());
event event_;
cl_int ret = clEnqueueCopyBufferRect(
m_queue,
src_buffer.get(),
dst_buffer.get(),
src_origin,
dst_origin,
region,
buffer_row_pitch,
buffer_slice_pitch,
host_row_pitch,
host_slice_pitch,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to unmap \p mem from the host memory space.
///
/// \see_opencl_ref{clEnqueueUnmapMemObject}
event enqueue_unmap_mem_object(cl_mem mem,
void *mapped_ptr,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
event event_;
cl_int ret = clEnqueueUnmapMemObject(
m_queue,
mem,
mapped_ptr,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Enqueues a command to write a rectangular region from host memory
/// to \p buffer.
///
/// \see_opencl_ref{clEnqueueWriteBufferRect}
///
/// \opencl_version_warning{1,1}
void enqueue_write_buffer_rect(const buffer &buffer,
const size_t buffer_origin[3],
const size_t host_origin[3],
const size_t region[3],
size_t buffer_row_pitch,
size_t buffer_slice_pitch,
size_t host_row_pitch,
size_t host_slice_pitch,
void *host_ptr,
const wait_list &events = wait_list())
{
BOOST_ASSERT(m_queue != 0);
BOOST_ASSERT(buffer.get_context() == this->get_context());
BOOST_ASSERT(host_ptr != 0);
cl_int ret = clEnqueueWriteBufferRect(
m_queue,
buffer.get(),
CL_TRUE,
buffer_origin,
host_origin,
region,
buffer_row_pitch,
buffer_slice_pitch,
host_row_pitch,
host_slice_pitch,
host_ptr,
events.size(),
events.get_event_ptr(),
0
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
}
/// Enqueues a command to copy \p size bytes of data from \p src_ptr to
/// \p dst_ptr. The operation is performed asynchronously.
///
/// \opencl_version_warning{2,0}
///
/// \see_opencl2_ref{clEnqueueSVMMemcpy}
event enqueue_svm_memcpy_async(void *dst_ptr,
const void *src_ptr,
size_t size,
const wait_list &events = wait_list())
{
event event_;
cl_int ret = clEnqueueSVMMemcpy(
m_queue,
CL_FALSE,
dst_ptr,
src_ptr,
size,
events.size(),
events.get_event_ptr(),
&event_.get()
);
if(ret != CL_SUCCESS){
BOOST_THROW_EXCEPTION(opencl_error(ret));
}
return event_;
}
/// Append wait list to wait list
inline void wait_list_append(wait_list &dst, const wait_list &src) {
dst.insert(dst.begin(), src.begin(), src.end());
}
/// Append event to wait list
inline void wait_list_append(wait_list &dst, const event &e) {
dst.push_back(e);
}
inline void insert_events_variadic(wait_list &l, Event&& event, Rest&&... rest)
{
l.insert(std::forward<Event>(event));
insert_events_variadic(l, std::forward<Rest>(rest)...);
}
