ViewVC Help
View File | Revision Log | Show Annotations | Download File
/cvs/OpenCL/OpenCL.pm
(Generate patch)

Comparing OpenCL/OpenCL.pm (file contents):
Revision 1.20 by root, Sun Nov 20 01:09:48 2011 UTC vs.
Revision 1.74 by root, Fri May 4 15:04:51 2012 UTC

… …
43 43
44OpenCL::Event objects are used to signal when something is complete. 44OpenCL::Event objects are used to signal when something is complete.
45 45
46=head2 HELPFUL RESOURCES 46=head2 HELPFUL RESOURCES
47 47
48The OpenCL spec used to develop this module (1.2 spec was available, but 48The OpenCL specs used to develop this module - download these and keept
49no implementation was available to me :). 49hema round, they are required reference material:
50 50
51 http://www.khronos.org/registry/cl/specs/opencl-1.1.pdf 51 http://www.khronos.org/registry/cl/specs/opencl-1.1.pdf
52 http://www.khronos.org/registry/cl/specs/opencl-1.2.pdf
53 http://www.khronos.org/registry/cl/specs/opencl-1.2-extensions.pdf
52 54
53OpenCL manpages: 55OpenCL manpages:
54 56
55 http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/ 57 http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/
58 http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/
56 59
57If you are into UML class diagrams, the following diagram might help - if 60If you are into UML class diagrams, the following diagram might help - if
58not, it will be mildly cobfusing: 61not, it will be mildly confusing (also, the class hierarchy of this module
62is much more fine-grained):
59 63
60 http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/classDiagram.html 64 http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/classDiagram.html
61 65
62Here's a tutorial from AMD (very AMD-centric, too), not sure how useful it 66Here's a tutorial from AMD (very AMD-centric, too), not sure how useful it
63is, but at least it's free of charge: 67is, but at least it's free of charge:
64 68
65 http://developer.amd.com/zones/OpenCLZone/courses/Documents/Introduction_to_OpenCL_Programming%20Training_Guide%20%28201005%29.pdf 69 http://developer.amd.com/zones/OpenCLZone/courses/Documents/Introduction_to_OpenCL_Programming%20Training_Guide%20%28201005%29.pdf
… …
101 105
102Best run this once to get a feel for the platforms and devices in your 106Best run this once to get a feel for the platforms and devices in your
103system. 107system.
104 108
105 for my $platform (OpenCL::platforms) { 109 for my $platform (OpenCL::platforms) {
106 printf "platform: %s\n", $platform->info (OpenCL::PLATFORM_NAME); 110 printf "platform: %s\n", $platform->name;
107 printf "extensions: %s\n", $platform->info (OpenCL::PLATFORM_EXTENSIONS); 111 printf "extensions: %s\n", $platform->extensions;
108 for my $device ($platform->devices) { 112 for my $device ($platform->devices) {
109 printf "+ device: %s\n", $device->info (OpenCL::DEVICE_NAME); 113 printf "+ device: %s\n", $device->name;
110 my $ctx = $device->context; 114 my $ctx = $platform->context (undef, [$device]);
111 # do stuff 115 # do stuff
112 } 116 }
113 } 117 }
114 118
115=head2 Get a useful context and a command queue. 119=head2 Get a useful context and a command queue.
… …
138=head2 Create a buffer with some predefined data, read it back synchronously, 142=head2 Create a buffer with some predefined data, read it back synchronously,
139then asynchronously. 143then asynchronously.
140 144
141 my $buf = $ctx->buffer_sv (OpenCL::MEM_COPY_HOST_PTR, "helmut"); 145 my $buf = $ctx->buffer_sv (OpenCL::MEM_COPY_HOST_PTR, "helmut");
142 146
143 $queue->enqueue_read_buffer ($buf, 1, 1, 3, my $data); 147 $queue->read_buffer ($buf, 1, 1, 3, my $data);
144 print "$data\n"; 148 print "$data\n";
145 149
146 my $ev = $queue->enqueue_read_buffer ($buf, 0, 1, 3, my $data); 150 my $ev = $queue->read_buffer ($buf, 0, 1, 3, my $data);
147 $ev->wait; 151 $ev->wait;
148 print "$data\n"; # prints "elm" 152 print "$data\n"; # prints "elm"
149 153
150=head2 Create and build a program, then create a kernel out of one of its 154=head2 Create and build a program, then create a kernel out of one of its
151functions. 155functions.
152 156
153 my $src = ' 157 my $src = '
154 __kernel void 158 kernel void
155 squareit (__global float *input, __global float *output) 159 squareit (global float *input, global float *output)
156 { 160 {
157 $id = get_global_id (0); 161 $id = get_global_id (0);
158 output [id] = input [id] * input [id]; 162 output [id] = input [id] * input [id];
159 } 163 }
160 '; 164 ';
161 165
162 my $prog = $ctx->program_with_source ($src); 166 my $prog = $ctx->build_program ($src);
163
164 # build croaks on compile errors, so catch it and print the compile errors
165 eval { $prog->build ($dev); 1 }
166 or die $prog->build_info ($dev, OpenCL::PROGRAM_BUILD_LOG);
167
168 my $kernel = $prog->kernel ("squareit"); 167 my $kernel = $prog->kernel ("squareit");
169 168
170=head2 Create some input and output float buffers, then call the 169=head2 Create some input and output float buffers, then call the
171'squareit' kernel on them. 170'squareit' kernel on them.
172 171
… …
176 # set buffer 175 # set buffer
177 $kernel->set_buffer (0, $input); 176 $kernel->set_buffer (0, $input);
178 $kernel->set_buffer (1, $output); 177 $kernel->set_buffer (1, $output);
179 178
180 # execute it for all 4 numbers 179 # execute it for all 4 numbers
181 $queue->enqueue_nd_range_kernel ($kernel, undef, [4], undef); 180 $queue->nd_range_kernel ($kernel, undef, [4], undef);
182 181
183 # enqueue a synchronous read 182 # enqueue a synchronous read
184 $queue->enqueue_read_buffer ($output, 1, 0, OpenCL::SIZEOF_FLOAT * 4, my $data); 183 $queue->read_buffer ($output, 1, 0, OpenCL::SIZEOF_FLOAT * 4, my $data);
185 184
186 # print the results: 185 # print the results:
187 printf "%s\n", join ", ", unpack "f*", $data; 186 printf "%s\n", join ", ", unpack "f*", $data;
188 187
189=head2 The same enqueue operations as before, but assuming an out-of-order queue, 188=head2 The same enqueue operations as before, but assuming an out-of-order queue,
190showing off barriers. 189showing off barriers.
191 190
192 # execute it for all 4 numbers 191 # execute it for all 4 numbers
193 $queue->enqueue_nd_range_kernel ($kernel, undef, [4], undef); 192 $queue->nd_range_kernel ($kernel, undef, [4], undef);
194 193
195 # enqueue a barrier to ensure in-order execution 194 # enqueue a barrier to ensure in-order execution
196 $queue->enqueue_barrier; 195 $queue->barrier;
197 196
198 # enqueue an async read 197 # enqueue an async read
199 $queue->enqueue_read_buffer ($output, 0, 0, OpenCL::SIZEOF_FLOAT * 4, my $data); 198 $queue->read_buffer ($output, 0, 0, OpenCL::SIZEOF_FLOAT * 4, my $data);
200 199
201 # wait for all requests to finish 200 # wait for all requests to finish
202 $queue->finish; 201 $queue->finish;
203 202
204=head2 The same enqueue operations as before, but assuming an out-of-order queue, 203=head2 The same enqueue operations as before, but assuming an out-of-order queue,
205showing off event objects and wait lists. 204showing off event objects and wait lists.
206 205
207 # execute it for all 4 numbers 206 # execute it for all 4 numbers
208 my $ev = $queue->enqueue_nd_range_kernel ($kernel, undef, [4], undef); 207 my $ev = $queue->nd_range_kernel ($kernel, undef, [4], undef);
209 208
210 # enqueue an async read 209 # enqueue an async read
211 $ev = $queue->enqueue_read_buffer ($output, 0, 0, OpenCL::SIZEOF_FLOAT * 4, my $data, $ev); 210 $ev = $queue->read_buffer ($output, 0, 0, OpenCL::SIZEOF_FLOAT * 4, my $data, $ev);
212 211
213 # wait for the last event to complete 212 # wait for the last event to complete
214 $ev->wait; 213 $ev->wait;
214
215=head2 Use the OpenGL module to share a texture between OpenCL and OpenGL and draw some julia
216set flight effect.
217
218This is quite a long example to get you going - you can download it from
219L<http://cvs.schmorp.de/OpenCL/examples/juliaflight>.
220
221 use OpenGL ":all";
222 use OpenCL;
223
224 my $S = $ARGV[0] || 256; # window/texture size, smaller is faster
225
226 # open a window and create a gl texture
227 OpenGL::glpOpenWindow width => $S, height => $S;
228 my $texid = glGenTextures_p 1;
229 glBindTexture GL_TEXTURE_2D, $texid;
230 glTexImage2D_c GL_TEXTURE_2D, 0, GL_RGBA8, $S, $S, 0, GL_RGBA, GL_UNSIGNED_BYTE, 0;
231
232 # find and use the first opencl device that let's us get a shared opengl context
233 my $platform;
234 my $dev;
235 my $ctx;
236
237 for (OpenCL::platforms) {
238 $platform = $_;
239 for ($platform->devices) {
240 $dev = $_;
241 $ctx = $platform->context ([OpenCL::GLX_DISPLAY_KHR, undef, OpenCL::GL_CONTEXT_KHR, undef], [$dev])
242 and last;
243 }
244 }
245
246 $ctx
247 or die "cannot find suitable OpenCL device\n";
248
249 my $queue = $ctx->queue ($dev);
250
251 # now attach an opencl image2d object to the opengl texture
252 my $tex = $ctx->gl_texture2d (OpenCL::MEM_WRITE_ONLY, GL_TEXTURE_2D, 0, $texid);
253
254 # now the boring opencl code
255 my $src = <<EOF;
256 kernel void
257 juliatunnel (write_only image2d_t img, float time)
258 {
259 int2 xy = (int2)(get_global_id (0), get_global_id (1));
260 float2 p = convert_float2 (xy) / $S.f * 2.f - 1.f;
261
262 float2 m = (float2)(1.f, p.y) / fabs (p.x); // tunnel
263 m.x = fabs (fmod (m.x + time * 0.05f, 4.f) - 2.f);
264
265 float2 z = m;
266 float2 c = (float2)(sin (time * 0.01133f), cos (time * 0.02521f));
267
268 for (int i = 0; i < 25 && dot (z, z) < 4.f; ++i) // standard julia
269 z = (float2)(z.x * z.x - z.y * z.y, 2.f * z.x * z.y) + c;
270
271 float3 colour = (float3)(z.x, z.y, atan2 (z.y, z.x));
272 write_imagef (img, xy, (float4)(colour * p.x * p.x, 1.));
273 }
274 EOF
275
276 my $prog = $ctx->build_program ($src);
277 my $kernel = $prog->kernel ("juliatunnel");
278
279 # program compiled, kernel ready, now draw and loop
280
281 for (my $time; ; ++$time) {
282 # acquire objects from opengl
283 $queue->acquire_gl_objects ([$tex]);
284
285 # configure and run our kernel
286 $kernel->setf ("mf", $tex, $time*2); # mf = memory object, float
287 $queue->nd_range_kernel ($kernel, undef, [$S, $S], undef);
288
289 # release objects to opengl again
290 $queue->release_gl_objects ([$tex]);
291
292 # wait
293 $queue->finish;
294
295 # now draw the texture, the defaults should be all right
296 glTexParameterf GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST;
297
298 glEnable GL_TEXTURE_2D;
299 glBegin GL_QUADS;
300 glTexCoord2f 0, 1; glVertex3i -1, -1, -1;
301 glTexCoord2f 0, 0; glVertex3i 1, -1, -1;
302 glTexCoord2f 1, 0; glVertex3i 1, 1, -1;
303 glTexCoord2f 1, 1; glVertex3i -1, 1, -1;
304 glEnd;
305
306 glXSwapBuffers;
307
308 select undef, undef, undef, 1/60;
309 }
310
311=head2 How to modify the previous example to not rely on GL sharing.
312
313For those poor souls with only a sucky CPU OpenCL implementation, you
314currently have to read the image into some perl scalar, and then modify a
315texture or use glDrawPixels or so).
316
317First, when you don't need gl sharing, you can create the context much simpler:
318
319 $ctx = $platform->context (undef, [$dev])
320
321To use a texture, you would modify the above example by creating an
322OpenCL::Image manually instead of deriving it from a texture:
323
324 my $tex = $ctx->image2d (OpenCL::MEM_WRITE_ONLY, OpenCL::RGBA, OpenCL::UNORM_INT8, $S, $S);
325
326And in the draw loop, intead of acquire_gl_objects/release_gl_objects, you
327would read the image2d after the kernel has written it:
328
329 $queue->read_image ($tex, 0, 0, 0, 0, $S, $S, 1, 0, 0, my $data);
330
331And then you would upload the pixel data to the texture (or use glDrawPixels):
332
333 glTexSubImage2D_s GL_TEXTURE_2D, 0, 0, 0, $S, $S, GL_RGBA, GL_UNSIGNED_BYTE, $data;
334
335The fully modified example can be found at
336L<http://cvs.schmorp.de/OpenCL/examples/juliaflight-nosharing>.
337
338=head2 Julia sets look soooo 80ies.
339
340Then colour them differently, e.g. using orbit traps! Replace the loop and
341colour calculation from the previous examples by this:
342
343 float2 dm = (float2)(1.f, 1.f);
344
345 for (int i = 0; i < 25; ++i)
346 {
347 z = (float2)(z.x * z.x - z.y * z.y, 2.f * z.x * z.y) + c;
348 dm = fmin (dm, (float2)(fabs (dot (z, z) - 1.f), fabs (z.x - 1.f)));
349 }
350
351 float3 colour = (float3)(dm.x * dm.y, dm.x * dm.y, dm.x);
352
353Also try C<-10.f> instead of C<-1.f>.
215 354
216=head1 DOCUMENTATION 355=head1 DOCUMENTATION
217 356
218=head2 BASIC CONVENTIONS 357=head2 BASIC CONVENTIONS
219 358
… …
241=item * Structures are often specified by flattening out their components 380=item * Structures are often specified by flattening out their components
242as with short vectors, and returned as arrayrefs. 381as with short vectors, and returned as arrayrefs.
243 382
244=item * When enqueuing commands, the wait list is specified by adding 383=item * When enqueuing commands, the wait list is specified by adding
245extra arguments to the function - anywhere a C<$wait_events...> argument 384extra arguments to the function - anywhere a C<$wait_events...> argument
246is documented this can be any number of event objects. 385is documented this can be any number of event objects. As an extsnion
386implemented by this module, C<undef> values will be ignored in the event
387list.
247 388
248=item * When enqueuing commands, if the enqueue method is called in void 389=item * When enqueuing commands, if the enqueue method is called in void
249context, no event is created. In all other contexts an event is returned 390context, no event is created. In all other contexts an event is returned
250by the method. 391by the method.
251 392
… …
271 ulong IV - Q 412 ulong IV - Q
272 float NV float f 413 float NV float f
273 half IV ushort S 414 half IV ushort S
274 double NV double d 415 double NV double d
275 416
417=head2 GLX SUPPORT
418
419Due to the sad state that OpenGL support is in in Perl (mostly the OpenGL
420module, which has little to no documentation and has little to no support
421for glX), this module, as a special extension, treats context creation
422properties C<OpenCL::GLX_DISPLAY_KHR> and C<OpenCL::GL_CONTEXT_KHR>
423specially: If either or both of these are C<undef>, then the OpenCL
424module tries to dynamically resolve C<glXGetCurrentDisplay> and
425C<glXGetCurrentContext>, call these functions and use their return values
426instead.
427
428For this to work, the OpenGL library must be loaded, a GLX context must
429have been created and be made current, and C<dlsym> must be available and
430capable of finding the function via C<RTLD_DEFAULT>.
431
432=head2 EVENT SYSTEM
433
434OpenCL can generate a number of (potentially) asynchronous events, for
435example, after compiling a program, to signal a context-related error or,
436perhaps most important, to signal completion of queued jobs (by setting
437callbacks on OpenCL::Event objects).
438
439The OpenCL module converts all these callbacks into events - you can
440still register callbacks, but they are not executed when your OpenCL
441implementation calls the actual callback, but only later. Therefore, none
442of the limitations of OpenCL callbacks apply to the perl implementation:
443it is perfectly safe to make blocking operations from event callbacks, and
444enqueued operations don't need to be flushed.
445
446To facilitate this, this module maintains an event queue - each
447time an asynchronous event happens, it is queued, and perl will be
448interrupted. This is implemented via the L<Async::Interrupt> module. In
449addition, this module has L<AnyEvent> support, so it can seamlessly
450integrate itself into many event loops.
451
452Since L<Async::Interrupt> is a bit hard to understand, here are some case examples:
453
454=head3 Don't use callbacks.
455
456When your program never uses any callbacks, then there will never be any
457notifications you need to take care of, and therefore no need to worry
458about all this.
459
460You can achieve a great deal by explicitly waiting for events, or using
461barriers and flush calls. In many programs, there is no need at all to
462tinker with asynchronous events.
463
464=head3 Use AnyEvent
465
466This module automatically registers a watcher that invokes all outstanding
467event callbacks when AnyEvent is initialised (and block asynchronous
468interruptions). Using this mode of operations is the safest and most
469recommended one.
470
471To use this, simply use AnyEvent and this module normally, make sure you
472have an event loop running:
473
474 use Gtk2 -init;
475 use AnyEvent;
476
477 # initialise AnyEvent, by creating a watcher, or:
478 AnyEvent::detect;
479
480 my $e = $queue->marker;
481 $e->cb (sub {
482 warn "opencl is finished\n";
483 })
484
485 main Gtk2;
486
487Note that this module will not initialise AnyEvent for you. Before
488AnyEvent is initialised, the module will asynchronously interrupt perl
489instead. To avoid any surprises, it's best to explicitly initialise
490AnyEvent.
491
492You can temporarily enable asynchronous interruptions (see next paragraph)
493by calling C<$OpenCL::INTERRUPT->unblock> and disable them again by
494calling C<$OpenCL::INTERRUPT->block>.
495
496=head3 Let yourself be interrupted at any time
497
498This mode is the default unless AnyEvent is loaded and initialised. In
499this mode, OpenCL asynchronously interrupts a running perl program. The
500emphasis is on both I<asynchronously> and I<running> here.
501
502Asynchronously means that perl might execute your callbacks at any
503time. For example, in the following code (I<THAT YOU SHOULD NOT COPY>),
504the C<until> loop following the marker call will be interrupted by the
505callback:
506
507 my $e = $queue->marker;
508 my $flag;
509 $e->cb (sub { $flag = 1 });
510 1 until $flag;
511 # $flag is now 1
512
513The reason why you shouldn't blindly copy the above code is that
514busy waiting is a really really bad thing, and really really bad for
515performance.
516
517While at first this asynchronous business might look exciting, it can be
518really hard, because you need to be prepared for the callback code to be
519executed at any time, which limits the amount of things the callback code
520can do safely.
521
522This can be mitigated somewhat by using C<<
523$OpenCL::INTERRUPT->scope_block >> (see the L<Async::Interrupt>
524documentation for details).
525
526The other problem is that your program must be actively I<running> to be
527interrupted. When you calculate stuff, your program is running. When you
528hang in some C functions or other block execution (by calling C<sleep>,
529C<select>, running an event loop and so on), your program is waiting, not
530running.
531
532One way around that would be to attach a read watcher to your event loop,
533listening for events on C<< $OpenCL::INTERRUPT->pipe_fileno >>, using a
534dummy callback (C<sub { }>) to temporarily execute some perl code.
535
536That is then awfully close to using the built-in AnyEvent support above,
537though, so consider that one instead.
538
539=head3 Be creative
540
541OpenCL exports the L<Async::Interrupt> object it uses in the global
542variable C<$OpenCL::INTERRUPT>. You can configure it in any way you like.
543
544So if you want to feel like a real pro, err, wait, if you feel no risk
545menas no fun, you can experiment by implementing your own mode of
546operations.
547
548=cut
549
550package OpenCL;
551
552use common::sense;
553use Carp ();
554use Async::Interrupt ();
555
556our $POLL_FUNC; # set by XS
557
558BEGIN {
559 our $VERSION = '0.99';
560
561 require XSLoader;
562 XSLoader::load (__PACKAGE__, $VERSION);
563
564 @OpenCL::Platform::ISA =
565 @OpenCL::Device::ISA =
566 @OpenCL::Context::ISA =
567 @OpenCL::Queue::ISA =
568 @OpenCL::Memory::ISA =
569 @OpenCL::Sampler::ISA =
570 @OpenCL::Program::ISA =
571 @OpenCL::Kernel::ISA =
572 @OpenCL::Event::ISA = OpenCL::Object::;
573
574 @OpenCL::SubDevice::ISA = OpenCL::Device::;
575
576 @OpenCL::Buffer::ISA =
577 @OpenCL::Image::ISA = OpenCL::Memory::;
578
579 @OpenCL::BufferObj::ISA = OpenCL::Buffer::;
580
581 @OpenCL::Image2D::ISA =
582 @OpenCL::Image3D::ISA =
583 @OpenCL::Image2DArray::ISA =
584 @OpenCL::Image1D::ISA =
585 @OpenCL::Image1DArray::ISA =
586 @OpenCL::Image1DBuffer::ISA = OpenCL::Image::;
587
588 @OpenCL::UserEvent::ISA = OpenCL::Event::;
589
590 @OpenCL::MappedBuffer::ISA =
591 @OpenCL::MappedImage::ISA = OpenCL::Mapped::;
592}
593
276=head2 THE OpenCL PACKAGE 594=head2 THE OpenCL PACKAGE
277 595
278=over 4 596=over 4
279 597
280=item $int = OpenCL::errno 598=item $int = OpenCL::errno
281 599
282The last error returned by a function - it's only valid after an error occured 600The last error returned by a function - it's only valid after an error occured
283and before calling another OpenCL function. 601and before calling another OpenCL function.
284 602
285=item $str = OpenCL::err2str $errval 603=item $str = OpenCL::err2str [$errval]
286 604
287Comverts an error value into a human readable string. 605Converts an error value into a human readable string. IF no error value is
606given, then the last error will be used (as returned by OpenCL::errno).
288 607
289=item $str = OpenCL::enum2str $enum 608=item $str = OpenCL::enum2str $enum
290 609
291Converts most enum values (inof parameter names, image format constants, 610Converts most enum values (of parameter names, image format constants,
292object types, addressing and filter modes, command types etc.) into a 611object types, addressing and filter modes, command types etc.) into a
293human readbale string. When confronted with some random integer it can be 612human readable string. When confronted with some random integer it can be
294very helpful to pass it through this function to maybe get some readable 613very helpful to pass it through this function to maybe get some readable
295string out of it. 614string out of it.
296 615
297=item @platforms = OpenCL::platforms 616=item @platforms = OpenCL::platforms
298 617
299Returns all available OpenCL::Platform objects. 618Returns all available OpenCL::Platform objects.
300 619
301L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetPlatformIDs.html> 620L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetPlatformIDs.html>
302 621
303=item $ctx = OpenCL::context_from_type $properties, $type = OpenCL::DEVICE_TYPE_DEFAULT, $notify = undef 622=item $ctx = OpenCL::context_from_type $properties, $type = OpenCL::DEVICE_TYPE_DEFAULT, $callback->($err, $pvt) = $print_stderr
304 623
305Tries to create a context from a default device and platform - never worked for me. 624Tries to create a context from a default device and platform type - never worked for me.
625
626type: OpenCL::DEVICE_TYPE_DEFAULT, OpenCL::DEVICE_TYPE_CPU, OpenCL::DEVICE_TYPE_GPU,
627OpenCL::DEVICE_TYPE_ACCELERATOR, OpenCL::DEVICE_TYPE_CUSTOM, OpenCL::DEVICE_TYPE_ALL.
306 628
307L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContextFromType.html> 629L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContextFromType.html>
308 630
631=item $ctx = OpenCL::context $properties, \@devices, $callback->($err, $pvt) = $print_stderr)
632
633Create a new OpenCL::Context object using the given device object(s). This
634function isn't implemented yet, use C<< $platform->context >> instead.
635
636L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContext.html>
637
309=item OpenCL::wait_for_events $wait_events... 638=item OpenCL::wait_for_events $wait_events...
310 639
311Waits for all events to complete. 640Waits for all events to complete.
312 641
313L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clWaitForEvents.html> 642L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clWaitForEvents.html>
314 643
644=item OpenCL::poll
645
646Checks if there are any outstanding events (see L<EVENT SYSTEM>) and
647invokes their callbacks.
648
649=item $OpenCL::INTERRUPT
650
651The L<Async::Interrupt> object used to signal asynchronous events (see
652L<EVENT SYSTEM>).
653
654=cut
655
656our $INTERRUPT = new Async::Interrupt c_cb => [$POLL_FUNC, 0];
657
658&_eq_initialise ($INTERRUPT->signal_func);
659
660=item $OpenCL::WATCHER
661
662The L<AnyEvent> watcher object used to watch for asynchronous events (see
663L<EVENT SYSTEM>). This variable is C<undef> until L<AnyEvent> has been
664loaded I<and> initialised (e.g. by calling C<AnyEvent::detect>).
665
666=cut
667
668our $WATCHER;
669
670sub _init_anyevent {
671 $INTERRUPT->block;
672 $WATCHER = AE::io ($INTERRUPT->pipe_fileno, 0, sub { $INTERRUPT->handle });
673}
674
675if (defined $AnyEvent::MODEL) {
676 _init_anyevent;
677} else {
678 push @AnyEvent::post_detect, \&_init_anyevent;
679}
680
315=back 681=back
316 682
683=head2 THE OpenCL::Object CLASS
684
685This is the base class for all objects in the OpenCL module. The only
686method it implements is the C<id> method, which is only useful if you want
687to interface to OpenCL on the C level.
688
689=over 4
690
691=item $iv = $obj->id
692
693OpenCL objects are represented by pointers or integers on the C level. If
694you want to interface to an OpenCL object directly on the C level, then
695you need this value, which is returned by this method. You should use an
696C<IV> type in your code and cast that to the correct type.
697
698=cut
699
700sub OpenCL::Object::id {
701 ref $_[0] eq "SCALAR"
702 ? ${ $_[0] }
703 : $_[0][0]
704}
705
706=back
707
317=head2 THE OpenCL::Platform CLASS 708=head2 THE OpenCL::Platform CLASS
318 709
319=over 4 710=over 4
320 711
321=item @devices = $platform->devices ($type = OpenCL::DEVICE_TYPE_ALL) 712=item @devices = $platform->devices ($type = OpenCL::DEVICE_TYPE_ALL)
322 713
323Returns a list of matching OpenCL::Device objects. 714Returns a list of matching OpenCL::Device objects.
324 715
325=item $ctx = $platform->context_from_type ($properties, $type = OpenCL::DEVICE_TYPE_DEFAULT, $notify = undef) 716=item $ctx = $platform->context_from_type ($properties, $type = OpenCL::DEVICE_TYPE_DEFAULT, $callback->($err, $pvt) = $print_stderr)
326 717
327Tries to create a context. Never worked for me, and you need devices explitly anyway. 718Tries to create a context. Never worked for me, and you need devices explicitly anyway.
328 719
329L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContextFromType.html> 720L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContextFromType.html>
330 721
331=item $ctx = $device->context ($properties = undef, @$devices, $notify = undef) 722=item $ctx = $platform->context ($properties, \@devices, $callback->($err, $pvt) = $print_stderr)
332 723
333Create a new OpenCL::Context object using the given device object(s)- a 724Create a new OpenCL::Context object using the given device object(s)- a
334CL_CONTEXT_PLATFORM property is supplied automatically. 725CL_CONTEXT_PLATFORM property is supplied automatically.
335 726
336L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContext.html> 727L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateContext.html>
337 728
338=item $packed_value = $platform->info ($name) 729=item $packed_value = $platform->info ($name)
339 730
340Calls C<clGetPlatformInfo> and returns the packed, raw value - for 731Calls C<clGetPlatformInfo> and returns the packed, raw value - for
341strings, this will be the string, for other values you probably need to 732strings, this will be the string (possibly including terminating \0), for
342use the correct C<unpack>. 733other values you probably need to use the correct C<unpack>.
343 734
344It's best to avoid this method and use one of the predefined C<get_*> 735It's best to avoid this method and use one of the following convenience
345methods. 736wrappers.
346 737
347L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetPlatformInfo.html> 738L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetPlatformInfo.html>
348 739
740=item $platform->unload_compiler
741
742Attempts to unload the compiler for this platform, for endless
743profit. Does nothing on OpenCL 1.1.
744
745L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clUnloadPlatformCompiler.html>
746
349=for gengetinfo begin platform 747=for gengetinfo begin platform
350 748
351
352=item $string = $platform->profile 749=item $string = $platform->profile
353 750
354Calls C<clGetPlatformInfo> with C<CL_PLATFORM_PROFILE> and returns the result(s). 751Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_PROFILE> and returns the result.
355 752
356=item $string = $platform->version 753=item $string = $platform->version
357 754
358Calls C<clGetPlatformInfo> with C<CL_PLATFORM_VERSION> and returns the result(s). 755Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_VERSION> and returns the result.
359 756
360=item $string = $platform->name 757=item $string = $platform->name
361 758
362Calls C<clGetPlatformInfo> with C<CL_PLATFORM_NAME> and returns the result(s). 759Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_NAME> and returns the result.
363 760
364=item $string = $platform->vendor 761=item $string = $platform->vendor
365 762
366Calls C<clGetPlatformInfo> with C<CL_PLATFORM_VENDOR> and returns the result(s). 763Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_VENDOR> and returns the result.
367 764
368=item $string = $platform->extensions 765=item $string = $platform->extensions
369 766
370Calls C<clGetPlatformInfo> with C<CL_PLATFORM_EXTENSIONS> and returns the result(s). 767Calls C<clGetPlatformInfo> with C<OpenCL::PLATFORM_EXTENSIONS> and returns the result.
768
371=for gengetinfo end platform 769=for gengetinfo end platform
372 770
373=back 771=back
374 772
375=head2 THE OpenCL::Device CLASS 773=head2 THE OpenCL::Device CLASS
… …
380 778
381See C<< $platform->info >> for details. 779See C<< $platform->info >> for details.
382 780
383L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetDeviceInfo.html> 781L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetDeviceInfo.html>
384 782
783=item @devices = $device->sub_devices (\@properties)
784
785Creates OpencL::SubDevice objects by partitioning an existing device.
786
787L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateSubDevices.html>
788
789=for gengetinfo begin device
790
791=item $device_type = $device->type
792
793Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_TYPE> and returns the result.
794
795=item $uint = $device->vendor_id
796
797Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_VENDOR_ID> and returns the result.
798
799=item $uint = $device->max_compute_units
800
801Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_COMPUTE_UNITS> and returns the result.
802
803=item $uint = $device->max_work_item_dimensions
804
805Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WORK_ITEM_DIMENSIONS> and returns the result.
806
807=item $int = $device->max_work_group_size
808
809Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WORK_GROUP_SIZE> and returns the result.
810
811=item @ints = $device->max_work_item_sizes
812
813Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WORK_ITEM_SIZES> and returns the result.
814
815=item $uint = $device->preferred_vector_width_char
816
817Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_CHAR> and returns the result.
818
819=item $uint = $device->preferred_vector_width_short
820
821Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_SHORT> and returns the result.
822
823=item $uint = $device->preferred_vector_width_int
824
825Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_INT> and returns the result.
826
827=item $uint = $device->preferred_vector_width_long
828
829Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_LONG> and returns the result.
830
831=item $uint = $device->preferred_vector_width_float
832
833Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_FLOAT> and returns the result.
834
835=item $uint = $device->preferred_vector_width_double
836
837Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_DOUBLE> and returns the result.
838
839=item $uint = $device->max_clock_frequency
840
841Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_CLOCK_FREQUENCY> and returns the result.
842
843=item $bitfield = $device->address_bits
844
845Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_ADDRESS_BITS> and returns the result.
846
847=item $uint = $device->max_read_image_args
848
849Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_READ_IMAGE_ARGS> and returns the result.
850
851=item $uint = $device->max_write_image_args
852
853Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_WRITE_IMAGE_ARGS> and returns the result.
854
855=item $ulong = $device->max_mem_alloc_size
856
857Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_MEM_ALLOC_SIZE> and returns the result.
858
859=item $int = $device->image2d_max_width
860
861Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE2D_MAX_WIDTH> and returns the result.
862
863=item $int = $device->image2d_max_height
864
865Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE2D_MAX_HEIGHT> and returns the result.
866
867=item $int = $device->image3d_max_width
868
869Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE3D_MAX_WIDTH> and returns the result.
870
871=item $int = $device->image3d_max_height
872
873Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE3D_MAX_HEIGHT> and returns the result.
874
875=item $int = $device->image3d_max_depth
876
877Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE3D_MAX_DEPTH> and returns the result.
878
879=item $uint = $device->image_support
880
881Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_IMAGE_SUPPORT> and returns the result.
882
883=item $int = $device->max_parameter_size
884
885Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_PARAMETER_SIZE> and returns the result.
886
887=item $uint = $device->max_samplers
888
889Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_SAMPLERS> and returns the result.
890
891=item $uint = $device->mem_base_addr_align
892
893Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MEM_BASE_ADDR_ALIGN> and returns the result.
894
895=item $uint = $device->min_data_type_align_size
896
897Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MIN_DATA_TYPE_ALIGN_SIZE> and returns the result.
898
899=item $device_fp_config = $device->single_fp_config
900
901Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_SINGLE_FP_CONFIG> and returns the result.
902
903=item $device_mem_cache_type = $device->global_mem_cache_type
904
905Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_CACHE_TYPE> and returns the result.
906
907=item $uint = $device->global_mem_cacheline_size
908
909Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_CACHELINE_SIZE> and returns the result.
910
911=item $ulong = $device->global_mem_cache_size
912
913Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_CACHE_SIZE> and returns the result.
914
915=item $ulong = $device->global_mem_size
916
917Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_GLOBAL_MEM_SIZE> and returns the result.
918
919=item $ulong = $device->max_constant_buffer_size
920
921Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_CONSTANT_BUFFER_SIZE> and returns the result.
922
923=item $uint = $device->max_constant_args
924
925Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_MAX_CONSTANT_ARGS> and returns the result.
926
927=item $device_local_mem_type = $device->local_mem_type
928
929Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_LOCAL_MEM_TYPE> and returns the result.
930
931=item $ulong = $device->local_mem_size
932
933Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_LOCAL_MEM_SIZE> and returns the result.
934
935=item $boolean = $device->error_correction_support
936
937Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_ERROR_CORRECTION_SUPPORT> and returns the result.
938
939=item $int = $device->profiling_timer_resolution
940
941Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PROFILING_TIMER_RESOLUTION> and returns the result.
942
943=item $boolean = $device->endian_little
944
945Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_ENDIAN_LITTLE> and returns the result.
946
947=item $boolean = $device->available
948
949Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_AVAILABLE> and returns the result.
950
951=item $boolean = $device->compiler_available
952
953Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_COMPILER_AVAILABLE> and returns the result.
954
955=item $device_exec_capabilities = $device->execution_capabilities
956
957Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_EXECUTION_CAPABILITIES> and returns the result.
958
959=item $command_queue_properties = $device->properties
960
961Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_QUEUE_PROPERTIES> and returns the result.
962
963=item $ = $device->platform
964
965Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PLATFORM> and returns the result.
966
967=item $string = $device->name
968
969Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NAME> and returns the result.
970
971=item $string = $device->vendor
972
973Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_VENDOR> and returns the result.
974
975=item $string = $device->driver_version
976
977Calls C<clGetDeviceInfo> with C<OpenCL::DRIVER_VERSION> and returns the result.
978
979=item $string = $device->profile
980
981Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PROFILE> and returns the result.
982
983=item $string = $device->version
984
985Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_VERSION> and returns the result.
986
987=item $string = $device->extensions
988
989Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_EXTENSIONS> and returns the result.
990
991=item $uint = $device->preferred_vector_width_half
992
993Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PREFERRED_VECTOR_WIDTH_HALF> and returns the result.
994
995=item $uint = $device->native_vector_width_char
996
997Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_CHAR> and returns the result.
998
999=item $uint = $device->native_vector_width_short
1000
1001Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_SHORT> and returns the result.
1002
1003=item $uint = $device->native_vector_width_int
1004
1005Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_INT> and returns the result.
1006
1007=item $uint = $device->native_vector_width_long
1008
1009Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_LONG> and returns the result.
1010
1011=item $uint = $device->native_vector_width_float
1012
1013Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_FLOAT> and returns the result.
1014
1015=item $uint = $device->native_vector_width_double
1016
1017Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_DOUBLE> and returns the result.
1018
1019=item $uint = $device->native_vector_width_half
1020
1021Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_NATIVE_VECTOR_WIDTH_HALF> and returns the result.
1022
1023=item $device_fp_config = $device->double_fp_config
1024
1025Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_DOUBLE_FP_CONFIG> and returns the result.
1026
1027=item $device_fp_config = $device->half_fp_config
1028
1029Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_HALF_FP_CONFIG> and returns the result.
1030
1031=item $boolean = $device->host_unified_memory
1032
1033Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_HOST_UNIFIED_MEMORY> and returns the result.
1034
1035=item $device = $device->parent_device_ext
1036
1037Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PARENT_DEVICE_EXT> and returns the result.
1038
1039=item @device_partition_property_exts = $device->partition_types_ext
1040
1041Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PARTITION_TYPES_EXT> and returns the result.
1042
1043=item @device_partition_property_exts = $device->affinity_domains_ext
1044
1045Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_AFFINITY_DOMAINS_EXT> and returns the result.
1046
1047=item $uint = $device->reference_count_ext
1048
1049Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_REFERENCE_COUNT_EXT> and returns the result.
1050
1051=item @device_partition_property_exts = $device->partition_style_ext
1052
1053Calls C<clGetDeviceInfo> with C<OpenCL::DEVICE_PARTITION_STYLE_EXT> and returns the result.
1054
1055=for gengetinfo end device
1056
385=back 1057=back
386 1058
387=head2 THE OpenCL::Context CLASS 1059=head2 THE OpenCL::Context CLASS
388 1060
1061An OpenCL::Context is basically a container, or manager, for a number of
1062devices of a platform. It is used to create all sorts of secondary objects
1063such as buffers, queues, programs and so on.
1064
1065All context creation functions and methods take a list of properties
1066(type-value pairs). All property values can be specified as integers -
1067some additionally support other types:
1068
389=over 4 1069=over 4
390 1070
1071=item OpenCL::CONTEXT_PLATFORM
1072
1073Also accepts OpenCL::Platform objects.
1074
1075=item OpenCL::GLX_DISPLAY_KHR
1076
1077Also accepts C<undef>, in which case a deep and troubling hack is engaged
1078to find the current glx display (see L<GLX SUPPORT>).
1079
1080=item OpenCL::GL_CONTEXT_KHR
1081
1082Also accepts C<undef>, in which case a deep and troubling hack is engaged
1083to find the current glx context (see L<GLX SUPPORT>).
1084
1085=back
1086
1087=over 4
1088
1089=item $prog = $ctx->build_program ($program, $options = "")
1090
1091This convenience function tries to build the program on all devices in
1092the context. If the build fails, then the function will C<croak> with the
1093build log. Otherwise ti returns the program object.
1094
1095The C<$program> can either be a C<OpenCL::Program> object or a string
1096containing the program. In the latter case, a program objetc will be
1097created automatically.
1098
1099=cut
1100
1101sub OpenCL::Context::build_program {
1102 my ($self, $prog, $options) = @_;
1103
1104 $prog = $self->program_with_source ($prog)
1105 unless ref $prog;
1106
1107 eval { $prog->build (undef, $options); 1 }
1108 or errno == BUILD_PROGRAM_FAILURE
1109 or errno == INVALID_BINARY # workaround nvidia bug
1110 or Carp::croak "OpenCL::Context->build_program: " . err2str;
1111
1112 # we check status for all devices
1113 for my $dev ($self->devices) {
1114 $prog->build_status ($dev) == BUILD_SUCCESS
1115 or Carp::croak "Building OpenCL program for device '" . $dev->name . "' failed:\n"
1116 . $prog->build_log ($dev);
1117 }
1118
1119 $prog
1120}
1121
391=item $queue = $ctx->queue ($device, $properties) 1122=item $queue = $ctx->queue ($device, $properties)
392 1123
393Create a new OpenCL::Queue object from the context and the given device. 1124Create a new OpenCL::Queue object from the context and the given device.
394 1125
395L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateCommandQueue.html> 1126L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateCommandQueue.html>
396 1127
1128Example: create an out-of-order queue.
1129
1130 $queue = $ctx->queue ($device, OpenCL::QUEUE_OUT_OF_ORDER_EXEC_MODE_ENABLE);
1131
397=item $ev = $ctx->user_event 1132=item $ev = $ctx->user_event
398 1133
399Creates a new OpenCL::UserEvent object. 1134Creates a new OpenCL::UserEvent object.
400 1135
401L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateUserEvent.html> 1136L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateUserEvent.html>
402 1137
403=item $buf = $ctx->buffer ($flags, $len) 1138=item $buf = $ctx->buffer ($flags, $len)
404 1139
405Creates a new OpenCL::Buffer object with the given flags and octet-size. 1140Creates a new OpenCL::Buffer (actually OpenCL::BufferObj) object with the
1141given flags and octet-size.
1142
1143flags: OpenCL::MEM_READ_WRITE, OpenCL::MEM_WRITE_ONLY, OpenCL::MEM_READ_ONLY,
1144OpenCL::MEM_USE_HOST_PTR, OpenCL::MEM_ALLOC_HOST_PTR, OpenCL::MEM_COPY_HOST_PTR,
1145OpenCL::MEM_HOST_WRITE_ONLY, OpenCL::MEM_HOST_READ_ONLY, OpenCL::MEM_HOST_NO_ACCESS.
406 1146
407L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateBuffer.html> 1147L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateBuffer.html>
408 1148
409=item $buf = $ctx->buffer_sv ($flags, $data) 1149=item $buf = $ctx->buffer_sv ($flags, $data)
410 1150
411Creates a new OpenCL::Buffer object and initialise it with the given data values. 1151Creates a new OpenCL::Buffer (actually OpenCL::BufferObj) object and
1152initialise it with the given data values.
1153
1154=item $img = $ctx->image ($self, $flags, $channel_order, $channel_type, $type, $width, $height, $depth = 0, $array_size = 0, $row_pitch = 0, $slice_pitch = 0, $num_mip_level = 0, $num_samples = 0, $*data = &PL_sv_undef)
1155
1156Creates a new OpenCL::Image object and optionally initialises it with
1157the given data values.
1158
1159channel_order: OpenCL::R, OpenCL::A, OpenCL::RG, OpenCL::RA, OpenCL::RGB,
1160OpenCL::RGBA, OpenCL::BGRA, OpenCL::ARGB, OpenCL::INTENSITY, OpenCL::LUMINANCE,
1161OpenCL::Rx, OpenCL::RGx, OpenCL::RGBx.
1162
1163channel_type: OpenCL::SNORM_INT8, OpenCL::SNORM_INT16, OpenCL::UNORM_INT8,
1164OpenCL::UNORM_INT16, OpenCL::UNORM_SHORT_565, OpenCL::UNORM_SHORT_555,
1165OpenCL::UNORM_INT_101010, OpenCL::SIGNED_INT8, OpenCL::SIGNED_INT16,
1166OpenCL::SIGNED_INT32, OpenCL::UNSIGNED_INT8, OpenCL::UNSIGNED_INT16,
1167OpenCL::UNSIGNED_INT32, OpenCL::HALF_FLOAT, OpenCL::FLOAT.
1168
1169
1170type: OpenCL::MEM_OBJECT_BUFFER, OpenCL::MEM_OBJECT_IMAGE2D,
1171OpenCL::MEM_OBJECT_IMAGE3D, OpenCL::MEM_OBJECT_IMAGE2D_ARRAY,
1172OpenCL::MEM_OBJECT_IMAGE1D, OpenCL::MEM_OBJECT_IMAGE1D_ARRAY,
1173OpenCL::MEM_OBJECT_IMAGE1D_BUFFER.
1174
1175L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateImage.html>
412 1176
413=item $img = $ctx->image2d ($flags, $channel_order, $channel_type, $width, $height, $row_pitch = 0, $data = undef) 1177=item $img = $ctx->image2d ($flags, $channel_order, $channel_type, $width, $height, $row_pitch = 0, $data = undef)
414 1178
415Creates a new OpenCL::Image2D object and optionally initialises it with the given data values. 1179Creates a new OpenCL::Image2D object and optionally initialises it with
1180the given data values.
416 1181
417L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateImage2D.html> 1182L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateImage2D.html>
418 1183
419=item $img = $ctx->image3d ($flags, $channel_order, $channel_type, $width, $height, $depth, $row_pitch = 0, $slice_pitch = 0, $data = undef) 1184=item $img = $ctx->image3d ($flags, $channel_order, $channel_type, $width, $height, $depth, $row_pitch = 0, $slice_pitch = 0, $data = undef)
420 1185
421Creates a new OpenCL::Image3D object and optionally initialises it with the given data values. 1186Creates a new OpenCL::Image3D object and optionally initialises it with
1187the given data values.
422 1188
423L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateImage3D.html> 1189L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateImage3D.html>
1190
1191=item $buffer = $ctx->gl_buffer ($flags, $bufobj)
1192
1193Creates a new OpenCL::Buffer (actually OpenCL::BufferObj) object that refers to the given
1194OpenGL buffer object.
1195
1196flags: OpenCL::MEM_READ_WRITE, OpenCL::MEM_READ_ONLY, OpenCL::MEM_WRITE_ONLY.
1197
1198http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLBuffer.html
1199
1200=item $img = $ctx->gl_texture ($flags, $target, $miplevel, $texture)
1201
1202Creates a new OpenCL::Image object that refers to the given OpenGL
1203texture object or buffer.
1204
1205target: GL_TEXTURE_1D, GL_TEXTURE_1D_ARRAY, GL_TEXTURE_BUFFER,
1206GL_TEXTURE_2D, GL_TEXTURE_2D_ARRAY, GL_TEXTURE_3D,
1207GL_TEXTURE_CUBE_MAP_POSITIVE_X, GL_TEXTURE_CUBE_MAP_POSITIVE_Y,
1208GL_TEXTURE_CUBE_MAP_POSITIVE_Z, GL_TEXTURE_CUBE_MAP_NEGATIVE_X,
1209GL_TEXTURE_CUBE_MAP_NEGATIVE_Y, GL_TEXTURE_CUBE_MAP_NEGATIVE_Z,
1210GL_TEXTURE_RECTANGLE/GL_TEXTURE_RECTANGLE_ARB.
1211
1212http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateFromGLTexture.html
1213
1214=item $img = $ctx->gl_texture2d ($flags, $target, $miplevel, $texture)
1215
1216Creates a new OpenCL::Image2D object that refers to the given OpenGL
12172D texture object.
1218
1219http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLTexture2D.html
1220
1221=item $img = $ctx->gl_texture3d ($flags, $target, $miplevel, $texture)
1222
1223Creates a new OpenCL::Image3D object that refers to the given OpenGL
12243D texture object.
1225
1226http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLTexture3D.html
1227
1228=item $ctx->gl_renderbuffer ($flags, $renderbuffer)
1229
1230Creates a new OpenCL::Image2D object that refers to the given OpenGL
1231render buffer.
1232
1233http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateFromGLRenderbuffer.html
424 1234
425=item @formats = $ctx->supported_image_formats ($flags, $image_type) 1235=item @formats = $ctx->supported_image_formats ($flags, $image_type)
426 1236
427Returns a list of matching image formats - each format is an arrayref with 1237Returns a list of matching image formats - each format is an arrayref with
428two values, $channel_order and $channel_type, in it. 1238two values, $channel_order and $channel_type, in it.
… …
431 1241
432=item $sampler = $ctx->sampler ($normalized_coords, $addressing_mode, $filter_mode) 1242=item $sampler = $ctx->sampler ($normalized_coords, $addressing_mode, $filter_mode)
433 1243
434Creates a new OpenCL::Sampler object. 1244Creates a new OpenCL::Sampler object.
435 1245
1246addressing_mode: OpenCL::ADDRESS_NONE, OpenCL::ADDRESS_CLAMP_TO_EDGE,
1247OpenCL::ADDRESS_CLAMP, OpenCL::ADDRESS_REPEAT, OpenCL::ADDRESS_MIRRORED_REPEAT.
1248
1249filter_mode: OpenCL::FILTER_NEAREST, OpenCL::FILTER_LINEAR.
1250
436L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateSampler.html> 1251L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateSampler.html>
437 1252
438=item $program = $ctx->program_with_source ($string) 1253=item $program = $ctx->program_with_source ($string)
439 1254
440Creates a new OpenCL::Program object from the given source code. 1255Creates a new OpenCL::Program object from the given source code.
441 1256
442L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateProgramWithSource.html> 1257L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateProgramWithSource.html>
443 1258
1259=item ($program, \@status) = $ctx->program_with_binary (\@devices, \@binaries)
1260
1261Creates a new OpenCL::Program object from the given binaries.
1262
1263L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateProgramWithBinary.html>
1264
1265Example: clone an existing program object that contains a successfully
1266compiled program, no matter how useless this is.
1267
1268 my $clone = $ctx->program_with_binary ([$prog->devices], [$prog->binaries]);
1269
1270=item $program = $ctx->program_with_built_in_kernels (\@devices, $kernel_names)
1271
1272Creates a new OpenCL::Program object from the given built-in kernel names.
1273
1274L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCreateProgramWithBuiltInKernels.html>
1275
444=item $packed_value = $ctx->info ($name) 1276=item $packed_value = $ctx->info ($name)
445 1277
446See C<< $platform->info >> for details. 1278See C<< $platform->info >> for details.
447 1279
448L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetContextInfo.html> 1280L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetContextInfo.html>
449 1281
450=for gengetinfo begin context 1282=for gengetinfo begin context
451 1283
1284=item $uint = $context->reference_count
1285
1286Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_REFERENCE_COUNT> and returns the result.
1287
1288=item @devices = $context->devices
1289
1290Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_DEVICES> and returns the result.
1291
1292=item @property_ints = $context->properties
1293
1294Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_PROPERTIES> and returns the result.
1295
1296=item $uint = $context->num_devices
1297
1298Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_NUM_DEVICES> and returns the result.
1299
452=for gengetinfo end context 1300=for gengetinfo end context
453 1301
454=back 1302=back
455 1303
456=head2 THE OpenCL::Queue CLASS 1304=head2 THE OpenCL::Queue CLASS
457 1305
458An OpenCL::Queue represents an execution queue for OpenCL. You execute 1306An OpenCL::Queue represents an execution queue for OpenCL. You execute
459requests by calling their respective C<enqueue_xxx> method and waitinf for 1307requests by calling their respective method and waiting for it to complete
460it to complete in some way. 1308in some way.
461 1309
462All the enqueue methods return an event object that can be used to wait 1310Most methods that enqueue some request return an event object that can
463for completion, unless the method is called in void context, in which case 1311be used to wait for completion (optionally using a callback), unless
464no event object is created. 1312the method is called in void context, in which case no event object is
1313created.
465 1314
466They also allow you to specify any number of other event objects that this 1315They also allow you to specify any number of other event objects that this
467request has to wait for before it starts executing, by simply passing the 1316request has to wait for before it starts executing, by simply passing the
468event objects as extra parameters to the enqueue methods. 1317event objects as extra parameters to the enqueue methods. To simplify
1318program design, this module ignores any C<undef> values in the list of
1319events. This makes it possible to code operations such as this, without
1320having to put a valid event object into C<$event> first:
1321
1322 $event = $queue->xxx (..., $event);
469 1323
470Queues execute in-order by default, without any parallelism, so in most 1324Queues execute in-order by default, without any parallelism, so in most
471cases (i.e. you use only one queue) it's not necessary to wait for or 1325cases (i.e. you use only one queue) it's not necessary to wait for or
472create event objects. 1326create event objects, althoguh an our of order queue is often a bit
1327faster.
473 1328
474=over 4 1329=over 4
475 1330
476=item $packed_value = $ctx->info ($name)
477
478See C<< $platform->info >> for details.
479
480L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetCommandQueueInfo.html>
481
482=item $ev = $queue->enqueue_read_buffer ($buffer, $blocking, $offset, $len, $data, $wait_events...) 1331=item $ev = $queue->read_buffer ($buffer, $blocking, $offset, $len, $data, $wait_events...)
483 1332
484Reads data from buffer into the given string. 1333Reads data from buffer into the given string.
485 1334
486L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadBuffer.html> 1335L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadBuffer.html>
487 1336
488=item $ev = $queue->enqueue_write_buffer ($buffer, $blocking, $offset, $data, $wait_events...) 1337=item $ev = $queue->write_buffer ($buffer, $blocking, $offset, $data, $wait_events...)
489 1338
490Writes data to buffer from the given string. 1339Writes data to buffer from the given string.
491 1340
492L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteBuffer.html> 1341L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteBuffer.html>
493 1342
494=item $ev = $queue->enqueue_copy_buffer ($src, $dst, $src_offset, $dst_offset, $len, $wait_events...) 1343=item $ev = $queue->copy_buffer ($src, $dst, $src_offset, $dst_offset, $len, $wait_events...)
495 1344
496L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBuffer.html> 1345L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBuffer.html>
497 1346
1347=item $ev = $queue->read_buffer_rect (OpenCL::Memory buf, cl_bool blocking, $buf_x, $buf_y, $buf_z, $host_x, $host_y, $host_z, $width, $height, $depth, $buf_row_pitch, $buf_slice_pitch, $host_row_pitch, $host_slice_pitch, $data, $wait_events...)
1348
1349http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadBufferRect.html
1350
1351=item $ev = $queue->write_buffer_rect (OpenCL::Memory buf, cl_bool blocking, $buf_x, $buf_y, $buf_z, $host_x, $host_y, $host_z, $width, $height, $depth, $buf_row_pitch, $buf_slice_pitch, $host_row_pitch, $host_slice_pitch, $data, $wait_events...)
1352
1353http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteBufferRect.html
1354
1355=item $ev = $queue->copy_buffer_to_image ($src_buffer, $dst_image, $src_offset, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $wait_events...)
1356
1357L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferToImage.html>
1358
498=item $ev = $queue->enqueue_read_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...) 1359=item $ev = $queue->read_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...)
1360
1361C<$row_pitch> (and C<$slice_pitch>) can be C<0>, in which case the OpenCL
1362module uses the image width (and height) to supply default values.
499 1363
500L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadImage.html> 1364L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadImage.html>
501 1365
502=item $ev = $queue->enqueue_write_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...) 1366=item $ev = $queue->write_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...)
503 1367
1368C<$row_pitch> (and C<$slice_pitch>) can be C<0>, in which case the OpenCL
1369module uses the image width (and height) to supply default values.
504L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteImage.html> 1370L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteImage.html>
505 1371
1372=item $ev = $queue->copy_image ($src_image, $dst_image, $src_x, $src_y, $src_z, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $wait_events...)
1373
1374L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImage.html>
1375
1376=item $ev = $queue->copy_image_to_buffer ($src_image, $dst_image, $src_x, $src_y, $src_z, $width, $height, $depth, $dst_offset, $wait_events...)
1377
1378L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImageToBuffer.html>
1379
506=item $ev = $queue->enqueue_copy_buffer_rect ($src, $dst, $src_x, $src_y, $src_z, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $src_row_pitch, $src_slice_pitch, $dst_row_pitch, $dst_slice_pitch, $wait_event...) 1380=item $ev = $queue->copy_buffer_rect ($src, $dst, $src_x, $src_y, $src_z, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $src_row_pitch, $src_slice_pitch, $dst_row_pitch, $dst_slice_pitch, $wait_event...)
507 1381
508Yeah. 1382Yeah.
509 1383
510L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferRect.html>
511
512=item $ev = $queue->enqueue_copy_buffer_to_image ($src_buffer, $dst_image, $src_offset, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $wait_events...)
513
514L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferToImage.html>. 1384L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferToImage.html>.
515 1385
516=item $ev = $queue->enqueue_copy_image ($src_image, $dst_image, $src_x, $src_y, $src_z, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $wait_events...) 1386=item $ev = $queue->fill_buffer ($mem, $pattern, $offset, $size, ...)
517 1387
1388Fills the given buffer object with repeated applications of C<$pattern>,
1389starting at C<$offset> for C<$size> octets.
1390
1391L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueFillBuffer.html>
1392
1393=item $ev = $queue->fill_image ($img, $r, $g, $b, $a, $x, $y, $z, $width, $height, $depth, ...)
1394
1395Fills the given image area with the given rgba colour components. The
1396components are normally floating point values between C<0> and C<1>,
1397except when the image channel data type is a signe dor unsigned
1398unnormalised format, in which case the range is determined by the format.
1399
518L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImage.html> 1400L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueFillImage.html>
519 1401
520=item $ev = $queue->enqueue_copy_image_to_buffer ($src_image, $dst_image, $src_x, $src_y, $src_z, $width, $height, $depth, $dst_offset, $wait_events...)
521
522L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImageToBuffer.html>
523
524=item $ev = $queue->enqueue_task ($kernel, $wait_events...) 1402=item $ev = $queue->task ($kernel, $wait_events...)
525 1403
526L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueTask.html> 1404L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueTask.html>
527 1405
528=item $ev = $queue->enqueue_nd_range_kernel ($kernel, @$global_work_offset, @$global_work_size, @$local_work_size, $wait_events...) 1406=item $ev = $queue->nd_range_kernel ($kernel, \@global_work_offset, \@global_work_size, \@local_work_size, $wait_events...)
529 1407
530Enqueues a kernel execution. 1408Enqueues a kernel execution.
531 1409
532@$global_work_size must be specified as a reference to an array of 1410\@global_work_size must be specified as a reference to an array of
533integers specifying the work sizes (element counts). 1411integers specifying the work sizes (element counts).
534 1412
535@$global_work_offset must be either C<undef> (in which case all offsets 1413\@global_work_offset must be either C<undef> (in which case all offsets
536are C<0>), or a reference to an array of work offsets, with the same number 1414are C<0>), or a reference to an array of work offsets, with the same number
537of elements as @$global_work_size. 1415of elements as \@global_work_size.
538 1416
539@$local_work_size must be either C<undef> (in which case the 1417\@local_work_size must be either C<undef> (in which case the
540implementation is supposed to choose good local work sizes), or a 1418implementation is supposed to choose good local work sizes), or a
541reference to an array of local work sizes, with the same number of 1419reference to an array of local work sizes, with the same number of
542elements as @$global_work_size. 1420elements as \@global_work_size.
543 1421
544L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html> 1422L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html>
545 1423
546=item $ev = $queue->enqueue_marker 1424=item $ev = $queue->migrate_mem_objects (\@mem_objects, $flags, $wait_events...)
547 1425
1426Migrates a number of OpenCL::Memory objects to or from the device.
1427
1428flags: OpenCL::MIGRATE_MEM_OBJECT_HOST, OpenCL::MIGRATE_MEM_OBJECT_CONTENT_UNDEFINED
1429
1430L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueMigrateMemObjects.html>
1431
1432=item $ev = $queue->acquire_gl_objects ([object, ...], $wait_events...)
1433
1434Enqueues a list (an array-ref of OpenCL::Memory objects) to be acquired
1435for subsequent OpenCL usage.
1436
548L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMarker.html> 1437L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueAcquireGLObjects.html>
549 1438
1439=item $ev = $queue->release_gl_objects ([object, ...], $wait_events...)
1440
1441Enqueues a list (an array-ref of OpenCL::Memory objects) to be released
1442for subsequent OpenGL usage.
1443
1444L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReleaseGLObjects.html>
1445
550=item $ev = $queue->enqueue_wait_for_events ($wait_events...) 1446=item $ev = $queue->wait_for_events ($wait_events...)
551 1447
552L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWaitForEvents.html> 1448L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWaitForEvents.html>
553 1449
554=item $queue->enqueue_barrier 1450=item $ev = $queue->marker ($wait_events...)
555 1451
1452L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueMarkerWithWaitList.html>
1453
1454=item $ev = $queue->barrier ($wait_events...)
1455
556L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueBarrier.html> 1456L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueBarrierWithWaitList.html>
557 1457
558=item $queue->flush 1458=item $queue->flush
559 1459
560L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFlush.html> 1460L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFlush.html>
561 1461
562=item $queue->finish 1462=item $queue->finish
563 1463
564L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFinish.html> 1464L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFinish.html>
565 1465
1466=item $packed_value = $queue->info ($name)
1467
1468See C<< $platform->info >> for details.
1469
1470L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetCommandQueueInfo.html>
1471
1472=for gengetinfo begin command_queue
1473
1474=item $ctx = $command_queue->context
1475
1476Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_CONTEXT> and returns the result.
1477
1478=item $device = $command_queue->device
1479
1480Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_DEVICE> and returns the result.
1481
1482=item $uint = $command_queue->reference_count
1483
1484Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_REFERENCE_COUNT> and returns the result.
1485
1486=item $command_queue_properties = $command_queue->properties
1487
1488Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_PROPERTIES> and returns the result.
1489
1490=for gengetinfo end command_queue
1491
566=back 1492=back
567 1493
1494=head3 MEMORY MAPPED BUFFERS
1495
1496OpenCL allows you to map buffers and images to host memory (read: perl
1497scalars). This is done much like reading or copying a buffer, by enqueuing
1498a map or unmap operation on the command queue.
1499
1500The map operations return an C<OpenCL::Mapped> object - see L<THE
1501OpenCL::Mapped CLASS> section for details on what to do with these
1502objects.
1503
1504The object will be unmapped automatically when the mapped object is
1505destroyed (you can use a barrier to make sure the unmap has finished,
1506before using the buffer in a kernel), but you can also enqueue an unmap
1507operation manually.
1508
1509=over 4
1510
1511=item $mapped_buffer = $queue->map_buffer ($buf, $blocking=1, $map_flags=OpenCL::MAP_READ|OpenCL::MAP_WRITE, $offset=0, $size=undef, $wait_events...)
1512
1513Maps the given buffer into host memory and returns an
1514C<OpenCL::MappedBuffer> object. If C<$size> is specified as undef, then
1515the map will extend to the end of the buffer.
1516
1517map_flags: OpenCL::MAP_READ, OpenCL::MAP_WRITE, OpenCL::MAP_WRITE_INVALIDATE_REGION.
1518
1519L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMapBuffer.html>
1520
1521Example: map the buffer $buf fully and replace the first 4 bytes by "abcd", then unmap.
1522
1523 {
1524 my $mapped = $queue->map_buffer ($buf, 1, OpenCL::MAP_WRITE);
1525 substr $$mapped, 0, 4, "abcd";
1526 } # asynchronously unmap because $mapped is destroyed
1527
1528=item $mapped_image = $queue->map_image ($img, $blocking=1, $map_flags=OpenCL::MAP_READ|OpenCL::MAP_WRITE, $x=0, $y=0, $z=0, $width=undef, $height=undef, $depth=undef, $wait_events...)
1529
1530Maps the given image area into host memory and return an
1531C<OpenCL::MappedImage> object.
1532
1533If any of C<$width>, C<$height> and/or C<$depth> are C<undef> then they
1534will be replaced by the maximum possible value.
1535
1536L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMapImage.html>
1537
1538Example: map an image (with OpenCL::UNSIGNED_INT8 channel type) and set
1539the first channel of the leftmost column to 5, then explicitly unmap
1540it. You are not necessarily meant to do it this way, this example just
1541shows you the accessors to use :)
1542
1543 my $mapped = $queue->map_image ($image, 1, OpenCL::MAP_WRITE);
1544
1545 $mapped->set ($_ * $mapped->row_pitch, pack "C", 5)
1546 for 0..$image->height;
1547
1548 $mapped->unmap;.
1549 $mapped->wait; # only needed for out of order queues normally
1550
1551=item $ev = $queue->unmap ($mapped, $wait_events...)
1552
1553Unmaps the data from host memory. You must not call any methods that
1554modify the data, or modify the data scalar directly, after calling this
1555method.
1556
1557The mapped event object will always be passed as part of the
1558$wait_events. The mapped event object will be replaced by the new event
1559object that this request creates.
1560
1561=back
1562
568=head2 THE OpenCL::Memory CLASS 1563=head2 THE OpenCL::Memory CLASS
569 1564
570This the superclass of all memory objects - OpenCL::Buffer, OpenCL::Image, 1565This the superclass of all memory objects - OpenCL::Buffer, OpenCL::Image,
571OpenCL::Image2D and OpenCL::Image3D. The subclasses of this class 1566OpenCL::Image2D and OpenCL::Image3D.
572currently only exist to allow type-checking.
573 1567
574=over 4 1568=over 4
575 1569
576=item $packed_value = $memory->info ($name) 1570=item $packed_value = $memory->info ($name)
577 1571
578See C<< $platform->info >> for details. 1572See C<< $platform->info >> for details.
579 1573
580L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetMemObjectInfo.html> 1574L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetMemObjectInfo.html>
581 1575
1576=for gengetinfo begin mem
1577
1578=item $mem_object_type = $mem->type
1579
1580Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_TYPE> and returns the result.
1581
1582=item $mem_flags = $mem->flags
1583
1584Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_FLAGS> and returns the result.
1585
1586=item $int = $mem->size
1587
1588Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_SIZE> and returns the result.
1589
1590=item $ptr_value = $mem->host_ptr
1591
1592Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_HOST_PTR> and returns the result.
1593
1594=item $uint = $mem->map_count
1595
1596Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_MAP_COUNT> and returns the result.
1597
1598=item $uint = $mem->reference_count
1599
1600Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_REFERENCE_COUNT> and returns the result.
1601
1602=item $ctx = $mem->context
1603
1604Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_CONTEXT> and returns the result.
1605
1606=item $mem = $mem->associated_memobject
1607
1608Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_ASSOCIATED_MEMOBJECT> and returns the result.
1609
1610=item $int = $mem->offset
1611
1612Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_OFFSET> and returns the result.
1613
1614=for gengetinfo end mem
1615
1616=item ($type, $name) = $mem->gl_object_info
1617
1618Returns the OpenGL object type (e.g. OpenCL::GL_OBJECT_TEXTURE2D) and the
1619object "name" (e.g. the texture name) used to create this memory object.
1620
1621L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetGLObjectInfo.html>
1622
582=back 1623=back
583 1624
1625=head2 THE OpenCL::Buffer CLASS
1626
1627This is a subclass of OpenCL::Memory, and the superclass of
1628OpenCL::BufferObj. Its purpose is simply to distinguish between buffers
1629and sub-buffers.
1630
1631=head2 THE OpenCL::BufferObj CLASS
1632
1633This is a subclass of OpenCL::Buffer and thus OpenCL::Memory. It exists
1634because one cna create sub buffers of OpenLC::BufferObj objects, but not
1635sub buffers from these sub buffers.
1636
1637=over 4
1638
1639=item $subbuf = $buf_obj->sub_buffer_region ($flags, $origin, $size)
1640
1641Creates an OpenCL::Buffer objects from this buffer and returns it. The
1642C<buffer_create_type> is assumed to be C<CL_BUFFER_CREATE_TYPE_REGION>.
1643
1644L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateSubBuffer.html>
1645
1646=back
1647
584=head2 THE OpenCL::Image CLASS 1648=head2 THE OpenCL::Image CLASS
585 1649
586This is the superclass of all image objects - OpenCL::Image2D and OpenCL::Image3D. 1650This is the superclass of all image objects - OpenCL::Image1D,
1651OpenCL::Image1DArray, OpenCL::Image1DBuffer, OpenCL::Image2D,
1652OpenCL::Image2DArray and OpenCL::Image3D.
587 1653
588=over 4 1654=over 4
589 1655
590=item $packed_value = $ev->image_info ($name) 1656=item $packed_value = $image->image_info ($name)
591 1657
592See C<< $platform->info >> for details. 1658See C<< $platform->info >> for details.
593 1659
594The reason this method is not called C<info> is that there already is an 1660The reason this method is not called C<info> is that there already is an
595C<< ->info >> method inherited from C<OpenCL::Memory>. 1661C<< ->info >> method inherited from C<OpenCL::Memory>.
596 1662
597L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetImageInfo.html> 1663L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetImageInfo.html>
598 1664
1665=item ($channel_order, $channel_data_type) = $image->format
1666
1667Returns the channel order and type used to create the image by calling
1668C<clGetImageInfo> with C<CL_IMAGE_FORMAT>.
1669
1670=for gengetinfo begin image
1671
1672=item $int = $image->element_size
1673
1674Calls C<clGetImageInfo> with C<OpenCL::IMAGE_ELEMENT_SIZE> and returns the result.
1675
1676=item $int = $image->row_pitch
1677
1678Calls C<clGetImageInfo> with C<OpenCL::IMAGE_ROW_PITCH> and returns the result.
1679
1680=item $int = $image->slice_pitch
1681
1682Calls C<clGetImageInfo> with C<OpenCL::IMAGE_SLICE_PITCH> and returns the result.
1683
1684=item $int = $image->width
1685
1686Calls C<clGetImageInfo> with C<OpenCL::IMAGE_WIDTH> and returns the result.
1687
1688=item $int = $image->height
1689
1690Calls C<clGetImageInfo> with C<OpenCL::IMAGE_HEIGHT> and returns the result.
1691
1692=item $int = $image->depth
1693
1694Calls C<clGetImageInfo> with C<OpenCL::IMAGE_DEPTH> and returns the result.
1695
1696=for gengetinfo end image
1697
1698=for gengetinfo begin gl_texture
1699
1700=item $GLenum = $gl_texture->target
1701
1702Calls C<clGetGLTextureInfo> with C<OpenCL::GL_TEXTURE_TARGET> and returns the result.
1703
1704=item $GLint = $gl_texture->gl_mipmap_level
1705
1706Calls C<clGetGLTextureInfo> with C<OpenCL::GL_MIPMAP_LEVEL> and returns the result.
1707
1708=for gengetinfo end gl_texture
1709
599=back 1710=back
600 1711
601=head2 THE OpenCL::Sampler CLASS 1712=head2 THE OpenCL::Sampler CLASS
602 1713
603=over 4 1714=over 4
… …
606 1717
607See C<< $platform->info >> for details. 1718See C<< $platform->info >> for details.
608 1719
609L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetSamplerInfo.html> 1720L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetSamplerInfo.html>
610 1721
1722=for gengetinfo begin sampler
1723
1724=item $uint = $sampler->reference_count
1725
1726Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_REFERENCE_COUNT> and returns the result.
1727
1728=item $ctx = $sampler->context
1729
1730Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_CONTEXT> and returns the result.
1731
1732=item $addressing_mode = $sampler->normalized_coords
1733
1734Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_NORMALIZED_COORDS> and returns the result.
1735
1736=item $filter_mode = $sampler->addressing_mode
1737
1738Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_ADDRESSING_MODE> and returns the result.
1739
1740=item $boolean = $sampler->filter_mode
1741
1742Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_FILTER_MODE> and returns the result.
1743
1744=for gengetinfo end sampler
1745
611=back 1746=back
612 1747
613=head2 THE OpenCL::Program CLASS 1748=head2 THE OpenCL::Program CLASS
614 1749
615=over 4 1750=over 4
616 1751
617=item $packed_value = $program->info ($name) 1752=item $program->build (\@devices = undef, $options = "", $cb->($program) = undef)
618 1753
619See C<< $platform->info >> for details.
620
621L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
622
623=item $program->build ($device, $options = "")
624
625Tries to build the program with the givne options. 1754Tries to build the program with the given options. See also the
1755C<$ctx->build> convenience function.
1756
1757If a callback is specified, then it will be called when compilation is
1758finished. Note that many OpenCL implementations block your program while
1759compiling whether you use a callback or not. See C<build_async> if you
1760want to make sure the build is done in the background.
1761
1762Note that some OpenCL implementations act up badly, and don't call the
1763callback in some error cases (but call it in others). This implementation
1764assumes the callback will always be called, and leaks memory if this is
1765not so. So best make sure you don't pass in invalid values.
1766
1767Some implementations fail with C<OpenCL::INVALID_BINARY> when the
1768compilation state is successful but some later stage fails.
1769
1770options: C<-D name>, C<-D name=definition>, C<-I dir>,
1771C<-cl-single-precision-constant>, C<-cl-denorms-are-zero>,
1772C<-cl-fp32-correctly-rounded-divide-sqrt>, C<-cl-opt-disable>,
1773C<-cl-mad-enable>, C<-cl-no-signed-zeros>, C<-cl-unsafe-math-optimizations>,
1774C<-cl-finite-math-only>, C<-cl-fast-relaxed-math>,
1775C<-w>, C<-Werror>, C<-cl-std=CL1.1/CL1.2>, C<-cl-kernel-arg-info>,
1776C<-create-library>, C<-enable-link-options>.
626 1777
627L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clBuildProgram.html> 1778L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clBuildProgram.html>
1779
1780=item $program->build_async (\@devices = undef, $options = "", $cb->($program) = undef)
1781
1782Similar to C<< ->build >>, except it starts a thread, and never fails (you
1783need to check the compilation status form the callback, or by polling).
1784
1785build_status: OpenCL::BUILD_SUCCESS, OpenCL::BUILD_NONE,
1786OpenCL::BUILD_ERROR, OpenCL::BUILD_IN_PROGRESS.
628 1787
629=item $packed_value = $program->build_info ($device, $name) 1788=item $packed_value = $program->build_info ($device, $name)
630 1789
631Similar to C<< $platform->info >>, but returns build info for a previous 1790Similar to C<< $platform->info >>, but returns build info for a previous
632build attempt for the given device. 1791build attempt for the given device.
… …
638Creates an OpenCL::Kernel object out of the named C<__kernel> function in 1797Creates an OpenCL::Kernel object out of the named C<__kernel> function in
639the program. 1798the program.
640 1799
641L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateKernel.html> 1800L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateKernel.html>
642 1801
1802=item @kernels = $program->kernels_in_program
1803
1804Returns all kernels successfully compiled for all devices in program.
1805
1806http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateKernelsInProgram.html
1807
1808=for gengetinfo begin program_build
1809
1810=item $build_status = $program->build_status ($device)
1811
1812Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_STATUS> and returns the result.
1813
1814=item $string = $program->build_options ($device)
1815
1816Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_OPTIONS> and returns the result.
1817
1818=item $string = $program->build_log ($device)
1819
1820Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_LOG> and returns the result.
1821
1822=for gengetinfo end program_build
1823
1824=item $packed_value = $program->info ($name)
1825
1826See C<< $platform->info >> for details.
1827
1828L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
1829
1830=for gengetinfo begin program
1831
1832=item $uint = $program->reference_count
1833
1834Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_REFERENCE_COUNT> and returns the result.
1835
1836=item $ctx = $program->context
1837
1838Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_CONTEXT> and returns the result.
1839
1840=item $uint = $program->num_devices
1841
1842Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_NUM_DEVICES> and returns the result.
1843
1844=item @devices = $program->devices
1845
1846Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_DEVICES> and returns the result.
1847
1848=item $string = $program->source
1849
1850Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_SOURCE> and returns the result.
1851
1852=item @ints = $program->binary_sizes
1853
1854Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_BINARY_SIZES> and returns the result.
1855
1856=for gengetinfo end program
1857
1858=item @blobs = $program->binaries
1859
1860Returns a string for the compiled binary for every device associated with
1861the program, empty strings indicate missing programs, and an empty result
1862means no program binaries are available.
1863
1864These "binaries" are often, in fact, informative low-level assembly
1865sources.
1866
1867L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
1868
643=back 1869=back
644 1870
645=head2 THE OpenCL::Kernel CLASS 1871=head2 THE OpenCL::Kernel CLASS
646 1872
647=over 4 1873=over 4
… …
650 1876
651See C<< $platform->info >> for details. 1877See C<< $platform->info >> for details.
652 1878
653L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelInfo.html> 1879L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelInfo.html>
654 1880
1881=for gengetinfo begin kernel
1882
1883=item $string = $kernel->function_name
1884
1885Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_FUNCTION_NAME> and returns the result.
1886
1887=item $uint = $kernel->num_args
1888
1889Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_NUM_ARGS> and returns the result.
1890
1891=item $uint = $kernel->reference_count
1892
1893Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_REFERENCE_COUNT> and returns the result.
1894
1895=item $ctx = $kernel->context
1896
1897Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_CONTEXT> and returns the result.
1898
1899=item $program = $kernel->program
1900
1901Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_PROGRAM> and returns the result.
1902
1903=for gengetinfo end kernel
1904
655=item $packed_value = $kernel->work_group_info ($device, $name) 1905=item $packed_value = $kernel->work_group_info ($device, $name)
1906
1907See C<< $platform->info >> for details.
1908
1909L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelWorkGroupInfo.html>
1910
1911=for gengetinfo begin kernel_work_group
1912
1913=item $int = $kernel->work_group_size ($device)
1914
1915Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_WORK_GROUP_SIZE> and returns the result.
1916
1917=item @ints = $kernel->compile_work_group_size ($device)
1918
1919Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_COMPILE_WORK_GROUP_SIZE> and returns the result.
1920
1921=item $ulong = $kernel->local_mem_size ($device)
1922
1923Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_LOCAL_MEM_SIZE> and returns the result.
1924
1925=item $int = $kernel->preferred_work_group_size_multiple ($device)
1926
1927Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE> and returns the result.
1928
1929=item $ulong = $kernel->private_mem_size ($device)
1930
1931Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_PRIVATE_MEM_SIZE> and returns the result.
1932
1933=for gengetinfo end kernel_work_group
1934
1935=item $packed_value = $kernel->arg_info ($idx, $name)
1936
1937See C<< $platform->info >> for details.
1938
1939L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clGetKernelArgInfo.html>
1940
1941=for gengetinfo begin kernel_arg
1942
1943=item $kernel_arg_address_qualifier = $kernel->arg_address_qualifier ($idx)
1944
1945Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_ADDRESS_QUALIFIER> and returns the result.
1946
1947=item $kernel_arg_access_qualifier = $kernel->arg_access_qualifier ($idx)
1948
1949Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_ACCESS_QUALIFIER> and returns the result.
1950
1951=item $string = $kernel->arg_type_name ($idx)
1952
1953Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_TYPE_NAME> and returns the result.
1954
1955=item $kernel_arg_type_qualifier = $kernel->arg_type_qualifier ($idx)
1956
1957Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_TYPE_QUALIFIER> and returns the result.
1958
1959=item $string = $kernel->arg_name ($idx)
1960
1961Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_NAME> and returns the result.
1962
1963=for gengetinfo end kernel_arg
1964
1965=item $kernel->setf ($format, ...)
1966
1967Sets the arguments of a kernel. Since OpenCL 1.1 doesn't have a generic
1968way to set arguments (and with OpenCL 1.2 it might be rather slow), you
1969need to specify a format argument, much as with C<printf>, to tell OpenCL
1970what type of argument it is.
1971
1972The format arguments are single letters:
1973
1974 c char
1975 C unsigned char
1976 s short
1977 S unsigned short
1978 i int
1979 I unsigned int
1980 l long
1981 L unsigned long
1982
1983 h half float (0..65535)
1984 f float
1985 d double
1986
1987 z local (octet size)
1988
1989 m memory object (buffer or image)
1990 a sampler
1991 e event
1992
1993Space characters in the format string are ignored.
1994
1995Example: set the arguments for a kernel that expects an int, two floats, a buffer and an image.
1996
1997 $kernel->setf ("i ff mm", 5, 0.5, 3, $buffer, $image);
1998
1999=item $kernel->set_TYPE ($index, $value)
2000
2001=item $kernel->set_char ($index, $value)
2002
2003=item $kernel->set_uchar ($index, $value)
2004
2005=item $kernel->set_short ($index, $value)
2006
2007=item $kernel->set_ushort ($index, $value)
2008
2009=item $kernel->set_int ($index, $value)
2010
2011=item $kernel->set_uint ($index, $value)
2012
2013=item $kernel->set_long ($index, $value)
2014
2015=item $kernel->set_ulong ($index, $value)
2016
2017=item $kernel->set_half ($index, $value)
2018
2019=item $kernel->set_float ($index, $value)
2020
2021=item $kernel->set_double ($index, $value)
2022
2023=item $kernel->set_memory ($index, $value)
2024
2025=item $kernel->set_buffer ($index, $value)
2026
2027=item $kernel->set_image ($index, $value)
2028
2029=item $kernel->set_sampler ($index, $value)
2030
2031=item $kernel->set_local ($index, $value)
2032
2033=item $kernel->set_event ($index, $value)
2034
2035This is a family of methods to set the kernel argument with the number
2036C<$index> to the give C<$value>.
2037
2038Chars and integers (including the half type) are specified as integers,
2039float and double as floating point values, memory/buffer/image must be
2040an object of that type or C<undef>, local-memory arguments are set by
2041specifying the size, and sampler and event must be objects of that type.
2042
2043Note that C<set_memory> works for all memory objects (all types of buffers
2044and images) - the main purpose of the more specific C<set_TYPE> functions
2045is type checking.
2046
2047Setting an argument for a kernel does NOT keep a reference to the object -
2048for example, if you set an argument to some image object, free the image,
2049and call the kernel, you will run into undefined behaviour.
2050
2051L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetKernelArg.html>
2052
2053=back
2054
2055=head2 THE OpenCL::Event CLASS
2056
2057This is the superclass for all event objects (including OpenCL::UserEvent
2058objects).
2059
2060=over 4
2061
2062=item $ev->wait
2063
2064Waits for the event to complete.
2065
2066L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clWaitForEvents.html>
2067
2068=item $ev->cb ($exec_callback_type, $callback->($event, $event_command_exec_status))
2069
2070Adds a callback to the callback stack for the given event type. There is
2071no way to remove a callback again.
2072
2073L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetEventCallback.html>
2074
2075=item $packed_value = $ev->info ($name)
2076
2077See C<< $platform->info >> for details.
2078
2079L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetEventInfo.html>
2080
2081=for gengetinfo begin event
2082
2083=item $queue = $event->command_queue
2084
2085Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_QUEUE> and returns the result.
2086
2087=item $command_type = $event->command_type
2088
2089Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_TYPE> and returns the result.
2090
2091=item $uint = $event->reference_count
2092
2093Calls C<clGetEventInfo> with C<OpenCL::EVENT_REFERENCE_COUNT> and returns the result.
2094
2095=item $uint = $event->command_execution_status
2096
2097Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_EXECUTION_STATUS> and returns the result.
2098
2099=item $ctx = $event->context
2100
2101Calls C<clGetEventInfo> with C<OpenCL::EVENT_CONTEXT> and returns the result.
2102
2103=for gengetinfo end event
2104
2105=item $packed_value = $ev->profiling_info ($name)
656 2106
657See C<< $platform->info >> for details. 2107See C<< $platform->info >> for details.
658 2108
659The reason this method is not called C<info> is that there already is an 2109The reason this method is not called C<info> is that there already is an
660C<< ->info >> method. 2110C<< ->info >> method.
661 2111
662L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelWorkGroupInfo.html>
663
664=item $kernel->set_TYPE ($index, $value)
665
666This is a family of methods to set the kernel argument with the number C<$index> to the give C<$value>.
667
668TYPE is one of C<char>, C<uchar>, C<short>, C<ushort>, C<int>, C<uint>,
669C<long>, C<ulong>, C<half>, C<float>, C<double>, C<memory>, C<buffer>,
670C<image2d>, C<image3d>, C<sampler> or C<event>.
671
672Chars and integers (including the half type) are specified as integers,
673float and double as floating point values, memory/buffer/image2d/image3d
674must be an object of that type or C<undef>, and sampler and event must be
675objects of that type.
676
677L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetKernelArg.html> 2112L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProfilingInfo.html>
2113
2114=for gengetinfo begin profiling
2115
2116=item $ulong = $event->profiling_command_queued
2117
2118Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_QUEUED> and returns the result.
2119
2120=item $ulong = $event->profiling_command_submit
2121
2122Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_SUBMIT> and returns the result.
2123
2124=item $ulong = $event->profiling_command_start
2125
2126Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_START> and returns the result.
2127
2128=item $ulong = $event->profiling_command_end
2129
2130Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_END> and returns the result.
2131
2132=for gengetinfo end profiling
678 2133
679=back 2134=back
680 2135
681=head2 THE OpenCL::Event CLASS 2136=head2 THE OpenCL::UserEvent CLASS
682 2137
683This is the superclass for all event objects (including OpenCL::UserEvent 2138This is a subclass of OpenCL::Event.
684objects).
685 2139
686=over 4 2140=over 4
687 2141
688=item $packed_value = $ev->info ($name) 2142=item $ev->set_status ($execution_status)
689 2143
690See C<< $platform->info >> for details. 2144Sets the execution status of the user event. Can only be called once,
2145either with OpenCL::COMPLETE or a negative number as status.
691 2146
692L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetEventInfo.html> 2147execution_status: OpenCL::COMPLETE or a negative integer.
693 2148
694=item $packed_value = $ev->profiling_info ($name)
695
696See C<< $platform->info >> for details.
697
698The reason this method is not called C<info> is that there already is an
699C<< ->info >> method.
700
701L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProfilingInfo.html>
702
703=item $ev->wait
704
705Waits for the event to complete.
706
707L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clWaitForEvents.html> 2149L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetUserEventStatus.html>
708 2150
709=back 2151=back
710 2152
711=head2 THE OpenCL::UserEvent CLASS 2153=head2 THE OpenCL::Mapped CLASS
712 2154
713This is a subclass of OpenCL::Event. 2155This class represents objects mapped into host memory. They are
2156represented by a blessed string scalar. The string data is the mapped
2157memory area, that is, if you read or write it, then the mapped object is
2158accessed directly.
2159
2160You must only ever use operations that modify the string in-place - for
2161example, a C<substr> that doesn't change the length, or maybe a regex that
2162doesn't change the length. Any other operation might cause the data to be
2163copied.
2164
2165When the object is destroyed it will enqueue an implicit unmap operation
2166on the queue that was used to create it.
2167
2168Keep in mind that you I<need> to unmap (or destroy) mapped objects before
2169OpenCL sees the changes, even if some implementations don't need this
2170sometimes.
2171
2172Example, replace the first two floats in the mapped buffer by 1 and 2.
2173
2174 my $mapped = $queue->map_buffer ($buf, ...
2175 $mapped->event->wait; # make sure it's there
2176
2177 # now replace first 8 bytes by new data, which is exactly 8 bytes long
2178 # we blindly assume device endianness to equal host endianness
2179 # (and of course, we assume iee 754 single precision floats :)
2180 substr $$mapped, 0, 8, pack "f*", 1, 2;
714 2181
715=over 4 2182=over 4
716 2183
717=item $ev->set_status ($execution_status) 2184=item $ev = $mapped->unmap ($wait_events...)
718 2185
719L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetUserEventStatus.html> 2186Unmaps the mapped memory object, using the queue originally used to create
2187it, quite similarly to C<< $queue->unmap ($mapped, ...) >>.
2188
2189=item $bool = $mapped->mapped
2190
2191Returns whether the object is still mapped - true before an C<unmap> is
2192enqueued, false afterwards.
2193
2194=item $ev = $mapped->event
2195
2196Return the event object associated with the mapped object. Initially, this
2197will be the event object created when mapping the object, and after an
2198unmap, this will be the event object that the unmap operation created.
2199
2200=item $mapped->wait
2201
2202Same as C<< $mapped->event->wait >> - makes sure no operations on this
2203mapped object are outstanding.
2204
2205=item $bytes = $mapped->size
2206
2207Returns the size of the mapped area, in bytes. Same as C<length $$mapped>.
2208
2209=item $ptr = $mapped->ptr
2210
2211Returns the raw memory address of the mapped area.
2212
2213=item $mapped->set ($offset, $data)
2214
2215Replaces the data at the given C<$offset> in the memory area by the new
2216C<$data>. This method is safer than direct manipulation of C<$mapped>
2217because it does bounds-checking, but also slower.
2218
2219=item $data = $mapped->get ($offset, $length)
2220
2221Returns (without copying) a scalar representing the data at the given
2222C<$offset> and C<$length> in the mapped memory area. This is the same as
2223the following substr, except much slower;
2224
2225 $data = substr $$mapped, $offset, $length
2226
2227=cut
2228
2229sub OpenCL::Mapped::get {
2230 substr ${$_[0]}, $_[1], $_[2]
2231}
720 2232
721=back 2233=back
722 2234
2235=head2 THE OpenCL::MappedBuffer CLASS
2236
2237This is a subclass of OpenCL::Mapped, representing mapped buffers.
2238
2239=head2 THE OpenCL::MappedImage CLASS
2240
2241This is a subclass of OpenCL::Mapped, representing mapped images.
2242
2243=over 4
2244
2245=item $bytes = $mapped->row_pitch
2246
2247=item $bytes = $mapped->slice_pitch
2248
2249Return the row or slice pitch of the image that has been mapped.
2250
2251=back
2252
2253
723=cut 2254=cut
724
725package OpenCL;
726
727use common::sense;
728
729BEGIN {
730 our $VERSION = '0.15';
731
732 require XSLoader;
733 XSLoader::load (__PACKAGE__, $VERSION);
734
735 @OpenCL::Buffer::ISA =
736 @OpenCL::Image::ISA = OpenCL::Memory::;
737
738 @OpenCL::Image2D::ISA =
739 @OpenCL::Image3D::ISA = OpenCL::Image::;
740
741 @OpenCL::UserEvent::ISA = OpenCL::Event::;
742}
743 2255
7441; 22561;
745 2257
746=head1 AUTHOR 2258=head1 AUTHOR
747 2259

Diff Legend

– Removed lines
+ Added lines
< Changed lines
> Changed lines