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.75 by root, Sat May 5 02:30:27 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
1276=item $program = $ctx->link_program (\@devices, $options, \@programs, $cb->($program) = undef)
1277
1278Links all (already compiled) program objects specified in C<@programs>
1279together and returns a new OpenCL::Program object with the result.
1280
1281L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clLinkProgram.html>
1282
444=item $packed_value = $ctx->info ($name) 1283=item $packed_value = $ctx->info ($name)
445 1284
446See C<< $platform->info >> for details. 1285See C<< $platform->info >> for details.
447 1286
448L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetContextInfo.html> 1287L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetContextInfo.html>
449 1288
450=for gengetinfo begin context 1289=for gengetinfo begin context
451 1290
1291=item $uint = $context->reference_count
1292
1293Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_REFERENCE_COUNT> and returns the result.
1294
1295=item @devices = $context->devices
1296
1297Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_DEVICES> and returns the result.
1298
1299=item @property_ints = $context->properties
1300
1301Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_PROPERTIES> and returns the result.
1302
1303=item $uint = $context->num_devices
1304
1305Calls C<clGetContextInfo> with C<OpenCL::CONTEXT_NUM_DEVICES> and returns the result.
1306
452=for gengetinfo end context 1307=for gengetinfo end context
453 1308
454=back 1309=back
455 1310
456=head2 THE OpenCL::Queue CLASS 1311=head2 THE OpenCL::Queue CLASS
457 1312
458An OpenCL::Queue represents an execution queue for OpenCL. You execute 1313An OpenCL::Queue represents an execution queue for OpenCL. You execute
459requests by calling their respective C<enqueue_xxx> method and waitinf for 1314requests by calling their respective method and waiting for it to complete
460it to complete in some way. 1315in some way.
461 1316
462All the enqueue methods return an event object that can be used to wait 1317Most methods that enqueue some request return an event object that can
463for completion, unless the method is called in void context, in which case 1318be used to wait for completion (optionally using a callback), unless
464no event object is created. 1319the method is called in void context, in which case no event object is
1320created.
465 1321
466They also allow you to specify any number of other event objects that this 1322They 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 1323request has to wait for before it starts executing, by simply passing the
468event objects as extra parameters to the enqueue methods. 1324event objects as extra parameters to the enqueue methods. To simplify
1325program design, this module ignores any C<undef> values in the list of
1326events. This makes it possible to code operations such as this, without
1327having to put a valid event object into C<$event> first:
1328
1329 $event = $queue->xxx (..., $event);
469 1330
470Queues execute in-order by default, without any parallelism, so in most 1331Queues 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 1332cases (i.e. you use only one queue) it's not necessary to wait for or
472create event objects. 1333create event objects, althoguh an our of order queue is often a bit
1334faster.
473 1335
474=over 4 1336=over 4
475 1337
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...) 1338=item $ev = $queue->read_buffer ($buffer, $blocking, $offset, $len, $data, $wait_events...)
483 1339
484Reads data from buffer into the given string. 1340Reads data from buffer into the given string.
485 1341
486L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadBuffer.html> 1342L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadBuffer.html>
487 1343
488=item $ev = $queue->enqueue_write_buffer ($buffer, $blocking, $offset, $data, $wait_events...) 1344=item $ev = $queue->write_buffer ($buffer, $blocking, $offset, $data, $wait_events...)
489 1345
490Writes data to buffer from the given string. 1346Writes data to buffer from the given string.
491 1347
492L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteBuffer.html> 1348L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteBuffer.html>
493 1349
494=item $ev = $queue->enqueue_copy_buffer ($src, $dst, $src_offset, $dst_offset, $len, $wait_events...) 1350=item $ev = $queue->copy_buffer ($src, $dst, $src_offset, $dst_offset, $len, $wait_events...)
495 1351
496L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBuffer.html> 1352L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBuffer.html>
497 1353
1354=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...)
1355
1356http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadBufferRect.html
1357
1358=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...)
1359
1360http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteBufferRect.html
1361
1362=item $ev = $queue->copy_buffer_to_image ($src_buffer, $dst_image, $src_offset, $dst_x, $dst_y, $dst_z, $width, $height, $depth, $wait_events...)
1363
1364L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferToImage.html>
1365
498=item $ev = $queue->enqueue_read_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...) 1366=item $ev = $queue->read_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...)
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.
499 1370
500L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadImage.html> 1371L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReadImage.html>
501 1372
502=item $ev = $queue->enqueue_write_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...) 1373=item $ev = $queue->write_image ($src, $blocking, $x, $y, $z, $width, $height, $depth, $row_pitch, $slice_pitch, $data, $wait_events...)
503 1374
1375C<$row_pitch> (and C<$slice_pitch>) can be C<0>, in which case the OpenCL
1376module 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> 1377L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWriteImage.html>
505 1378
1379=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...)
1380
1381L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImage.html>
1382
1383=item $ev = $queue->copy_image_to_buffer ($src_image, $dst_image, $src_x, $src_y, $src_z, $width, $height, $depth, $dst_offset, $wait_events...)
1384
1385L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImageToBuffer.html>
1386
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...) 1387=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 1388
508Yeah. 1389Yeah.
509 1390
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>. 1391L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyBufferToImage.html>.
515 1392
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...) 1393=item $ev = $queue->fill_buffer ($mem, $pattern, $offset, $size, ...)
517 1394
1395Fills the given buffer object with repeated applications of C<$pattern>,
1396starting at C<$offset> for C<$size> octets.
1397
1398L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueFillBuffer.html>
1399
1400=item $ev = $queue->fill_image ($img, $r, $g, $b, $a, $x, $y, $z, $width, $height, $depth, ...)
1401
1402Fills the given image area with the given rgba colour components. The
1403components are normally floating point values between C<0> and C<1>,
1404except when the image channel data type is a signe dor unsigned
1405unnormalised format, in which case the range is determined by the format.
1406
518L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueCopyImage.html> 1407L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueFillImage.html>
519 1408
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...) 1409=item $ev = $queue->task ($kernel, $wait_events...)
525 1410
526L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueTask.html> 1411L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueTask.html>
527 1412
528=item $ev = $queue->enqueue_nd_range_kernel ($kernel, @$global_work_offset, @$global_work_size, @$local_work_size, $wait_events...) 1413=item $ev = $queue->nd_range_kernel ($kernel, \@global_work_offset, \@global_work_size, \@local_work_size, $wait_events...)
529 1414
530Enqueues a kernel execution. 1415Enqueues a kernel execution.
531 1416
532@$global_work_size must be specified as a reference to an array of 1417\@global_work_size must be specified as a reference to an array of
533integers specifying the work sizes (element counts). 1418integers specifying the work sizes (element counts).
534 1419
535@$global_work_offset must be either C<undef> (in which case all offsets 1420\@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 1421are C<0>), or a reference to an array of work offsets, with the same number
537of elements as @$global_work_size. 1422of elements as \@global_work_size.
538 1423
539@$local_work_size must be either C<undef> (in which case the 1424\@local_work_size must be either C<undef> (in which case the
540implementation is supposed to choose good local work sizes), or a 1425implementation is supposed to choose good local work sizes), or a
541reference to an array of local work sizes, with the same number of 1426reference to an array of local work sizes, with the same number of
542elements as @$global_work_size. 1427elements as \@global_work_size.
543 1428
544L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html> 1429L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html>
545 1430
546=item $ev = $queue->enqueue_marker 1431=item $ev = $queue->migrate_mem_objects (\@mem_objects, $flags, $wait_events...)
547 1432
1433Migrates a number of OpenCL::Memory objects to or from the device.
1434
1435flags: OpenCL::MIGRATE_MEM_OBJECT_HOST, OpenCL::MIGRATE_MEM_OBJECT_CONTENT_UNDEFINED
1436
1437L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueMigrateMemObjects.html>
1438
1439=item $ev = $queue->acquire_gl_objects ([object, ...], $wait_events...)
1440
1441Enqueues a list (an array-ref of OpenCL::Memory objects) to be acquired
1442for subsequent OpenCL usage.
1443
548L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMarker.html> 1444L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueAcquireGLObjects.html>
549 1445
1446=item $ev = $queue->release_gl_objects ([object, ...], $wait_events...)
1447
1448Enqueues a list (an array-ref of OpenCL::Memory objects) to be released
1449for subsequent OpenGL usage.
1450
1451L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueReleaseGLObjects.html>
1452
550=item $ev = $queue->enqueue_wait_for_events ($wait_events...) 1453=item $ev = $queue->wait_for_events ($wait_events...)
551 1454
552L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWaitForEvents.html> 1455L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueWaitForEvents.html>
553 1456
554=item $queue->enqueue_barrier 1457=item $ev = $queue->marker ($wait_events...)
555 1458
1459L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueMarkerWithWaitList.html>
1460
1461=item $ev = $queue->barrier ($wait_events...)
1462
556L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueBarrier.html> 1463L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clEnqueueBarrierWithWaitList.html>
557 1464
558=item $queue->flush 1465=item $queue->flush
559 1466
560L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFlush.html> 1467L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFlush.html>
561 1468
562=item $queue->finish 1469=item $queue->finish
563 1470
564L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFinish.html> 1471L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clFinish.html>
565 1472
1473=item $packed_value = $queue->info ($name)
1474
1475See C<< $platform->info >> for details.
1476
1477L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetCommandQueueInfo.html>
1478
1479=for gengetinfo begin command_queue
1480
1481=item $ctx = $command_queue->context
1482
1483Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_CONTEXT> and returns the result.
1484
1485=item $device = $command_queue->device
1486
1487Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_DEVICE> and returns the result.
1488
1489=item $uint = $command_queue->reference_count
1490
1491Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_REFERENCE_COUNT> and returns the result.
1492
1493=item $command_queue_properties = $command_queue->properties
1494
1495Calls C<clGetCommandQueueInfo> with C<OpenCL::QUEUE_PROPERTIES> and returns the result.
1496
1497=for gengetinfo end command_queue
1498
566=back 1499=back
567 1500
1501=head3 MEMORY MAPPED BUFFERS
1502
1503OpenCL allows you to map buffers and images to host memory (read: perl
1504scalars). This is done much like reading or copying a buffer, by enqueuing
1505a map or unmap operation on the command queue.
1506
1507The map operations return an C<OpenCL::Mapped> object - see L<THE
1508OpenCL::Mapped CLASS> section for details on what to do with these
1509objects.
1510
1511The object will be unmapped automatically when the mapped object is
1512destroyed (you can use a barrier to make sure the unmap has finished,
1513before using the buffer in a kernel), but you can also enqueue an unmap
1514operation manually.
1515
1516=over 4
1517
1518=item $mapped_buffer = $queue->map_buffer ($buf, $blocking=1, $map_flags=OpenCL::MAP_READ|OpenCL::MAP_WRITE, $offset=0, $size=undef, $wait_events...)
1519
1520Maps the given buffer into host memory and returns an
1521C<OpenCL::MappedBuffer> object. If C<$size> is specified as undef, then
1522the map will extend to the end of the buffer.
1523
1524map_flags: OpenCL::MAP_READ, OpenCL::MAP_WRITE, OpenCL::MAP_WRITE_INVALIDATE_REGION.
1525
1526L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMapBuffer.html>
1527
1528Example: map the buffer $buf fully and replace the first 4 bytes by "abcd", then unmap.
1529
1530 {
1531 my $mapped = $queue->map_buffer ($buf, 1, OpenCL::MAP_WRITE);
1532 substr $$mapped, 0, 4, "abcd";
1533 } # asynchronously unmap because $mapped is destroyed
1534
1535=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...)
1536
1537Maps the given image area into host memory and return an
1538C<OpenCL::MappedImage> object.
1539
1540If any of C<$width>, C<$height> and/or C<$depth> are C<undef> then they
1541will be replaced by the maximum possible value.
1542
1543L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueMapImage.html>
1544
1545Example: map an image (with OpenCL::UNSIGNED_INT8 channel type) and set
1546the first channel of the leftmost column to 5, then explicitly unmap
1547it. You are not necessarily meant to do it this way, this example just
1548shows you the accessors to use :)
1549
1550 my $mapped = $queue->map_image ($image, 1, OpenCL::MAP_WRITE);
1551
1552 $mapped->set ($_ * $mapped->row_pitch, pack "C", 5)
1553 for 0..$image->height;
1554
1555 $mapped->unmap;.
1556 $mapped->wait; # only needed for out of order queues normally
1557
1558=item $ev = $queue->unmap ($mapped, $wait_events...)
1559
1560Unmaps the data from host memory. You must not call any methods that
1561modify the data, or modify the data scalar directly, after calling this
1562method.
1563
1564The mapped event object will always be passed as part of the
1565$wait_events. The mapped event object will be replaced by the new event
1566object that this request creates.
1567
1568=back
1569
568=head2 THE OpenCL::Memory CLASS 1570=head2 THE OpenCL::Memory CLASS
569 1571
570This the superclass of all memory objects - OpenCL::Buffer, OpenCL::Image, 1572This the superclass of all memory objects - OpenCL::Buffer, OpenCL::Image,
571OpenCL::Image2D and OpenCL::Image3D. The subclasses of this class 1573OpenCL::Image2D and OpenCL::Image3D.
572currently only exist to allow type-checking.
573 1574
574=over 4 1575=over 4
575 1576
576=item $packed_value = $memory->info ($name) 1577=item $packed_value = $memory->info ($name)
577 1578
578See C<< $platform->info >> for details. 1579See C<< $platform->info >> for details.
579 1580
580L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetMemObjectInfo.html> 1581L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetMemObjectInfo.html>
581 1582
1583=for gengetinfo begin mem
1584
1585=item $mem_object_type = $mem->type
1586
1587Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_TYPE> and returns the result.
1588
1589=item $mem_flags = $mem->flags
1590
1591Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_FLAGS> and returns the result.
1592
1593=item $int = $mem->size
1594
1595Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_SIZE> and returns the result.
1596
1597=item $ptr_value = $mem->host_ptr
1598
1599Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_HOST_PTR> and returns the result.
1600
1601=item $uint = $mem->map_count
1602
1603Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_MAP_COUNT> and returns the result.
1604
1605=item $uint = $mem->reference_count
1606
1607Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_REFERENCE_COUNT> and returns the result.
1608
1609=item $ctx = $mem->context
1610
1611Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_CONTEXT> and returns the result.
1612
1613=item $mem = $mem->associated_memobject
1614
1615Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_ASSOCIATED_MEMOBJECT> and returns the result.
1616
1617=item $int = $mem->offset
1618
1619Calls C<clGetMemObjectInfo> with C<OpenCL::MEM_OFFSET> and returns the result.
1620
1621=for gengetinfo end mem
1622
1623=item ($type, $name) = $mem->gl_object_info
1624
1625Returns the OpenGL object type (e.g. OpenCL::GL_OBJECT_TEXTURE2D) and the
1626object "name" (e.g. the texture name) used to create this memory object.
1627
1628L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetGLObjectInfo.html>
1629
582=back 1630=back
583 1631
1632=head2 THE OpenCL::Buffer CLASS
1633
1634This is a subclass of OpenCL::Memory, and the superclass of
1635OpenCL::BufferObj. Its purpose is simply to distinguish between buffers
1636and sub-buffers.
1637
1638=head2 THE OpenCL::BufferObj CLASS
1639
1640This is a subclass of OpenCL::Buffer and thus OpenCL::Memory. It exists
1641because one cna create sub buffers of OpenLC::BufferObj objects, but not
1642sub buffers from these sub buffers.
1643
1644=over 4
1645
1646=item $subbuf = $buf_obj->sub_buffer_region ($flags, $origin, $size)
1647
1648Creates an OpenCL::Buffer objects from this buffer and returns it. The
1649C<buffer_create_type> is assumed to be C<CL_BUFFER_CREATE_TYPE_REGION>.
1650
1651L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateSubBuffer.html>
1652
1653=back
1654
584=head2 THE OpenCL::Image CLASS 1655=head2 THE OpenCL::Image CLASS
585 1656
586This is the superclass of all image objects - OpenCL::Image2D and OpenCL::Image3D. 1657This is the superclass of all image objects - OpenCL::Image1D,
1658OpenCL::Image1DArray, OpenCL::Image1DBuffer, OpenCL::Image2D,
1659OpenCL::Image2DArray and OpenCL::Image3D.
587 1660
588=over 4 1661=over 4
589 1662
590=item $packed_value = $ev->image_info ($name) 1663=item $packed_value = $image->image_info ($name)
591 1664
592See C<< $platform->info >> for details. 1665See C<< $platform->info >> for details.
593 1666
594The reason this method is not called C<info> is that there already is an 1667The reason this method is not called C<info> is that there already is an
595C<< ->info >> method inherited from C<OpenCL::Memory>. 1668C<< ->info >> method inherited from C<OpenCL::Memory>.
596 1669
597L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetImageInfo.html> 1670L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetImageInfo.html>
598 1671
1672=item ($channel_order, $channel_data_type) = $image->format
1673
1674Returns the channel order and type used to create the image by calling
1675C<clGetImageInfo> with C<CL_IMAGE_FORMAT>.
1676
1677=for gengetinfo begin image
1678
1679=item $int = $image->element_size
1680
1681Calls C<clGetImageInfo> with C<OpenCL::IMAGE_ELEMENT_SIZE> and returns the result.
1682
1683=item $int = $image->row_pitch
1684
1685Calls C<clGetImageInfo> with C<OpenCL::IMAGE_ROW_PITCH> and returns the result.
1686
1687=item $int = $image->slice_pitch
1688
1689Calls C<clGetImageInfo> with C<OpenCL::IMAGE_SLICE_PITCH> and returns the result.
1690
1691=item $int = $image->width
1692
1693Calls C<clGetImageInfo> with C<OpenCL::IMAGE_WIDTH> and returns the result.
1694
1695=item $int = $image->height
1696
1697Calls C<clGetImageInfo> with C<OpenCL::IMAGE_HEIGHT> and returns the result.
1698
1699=item $int = $image->depth
1700
1701Calls C<clGetImageInfo> with C<OpenCL::IMAGE_DEPTH> and returns the result.
1702
1703=for gengetinfo end image
1704
1705=for gengetinfo begin gl_texture
1706
1707=item $GLenum = $gl_texture->target
1708
1709Calls C<clGetGLTextureInfo> with C<OpenCL::GL_TEXTURE_TARGET> and returns the result.
1710
1711=item $GLint = $gl_texture->gl_mipmap_level
1712
1713Calls C<clGetGLTextureInfo> with C<OpenCL::GL_MIPMAP_LEVEL> and returns the result.
1714
1715=for gengetinfo end gl_texture
1716
599=back 1717=back
600 1718
601=head2 THE OpenCL::Sampler CLASS 1719=head2 THE OpenCL::Sampler CLASS
602 1720
603=over 4 1721=over 4
… …
606 1724
607See C<< $platform->info >> for details. 1725See C<< $platform->info >> for details.
608 1726
609L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetSamplerInfo.html> 1727L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetSamplerInfo.html>
610 1728
1729=for gengetinfo begin sampler
1730
1731=item $uint = $sampler->reference_count
1732
1733Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_REFERENCE_COUNT> and returns the result.
1734
1735=item $ctx = $sampler->context
1736
1737Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_CONTEXT> and returns the result.
1738
1739=item $addressing_mode = $sampler->normalized_coords
1740
1741Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_NORMALIZED_COORDS> and returns the result.
1742
1743=item $filter_mode = $sampler->addressing_mode
1744
1745Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_ADDRESSING_MODE> and returns the result.
1746
1747=item $boolean = $sampler->filter_mode
1748
1749Calls C<clGetSamplerInfo> with C<OpenCL::SAMPLER_FILTER_MODE> and returns the result.
1750
1751=for gengetinfo end sampler
1752
611=back 1753=back
612 1754
613=head2 THE OpenCL::Program CLASS 1755=head2 THE OpenCL::Program CLASS
614 1756
615=over 4 1757=over 4
616 1758
617=item $packed_value = $program->info ($name) 1759=item $program->build (\@devices = undef, $options = "", $cb->($program) = undef)
618 1760
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. 1761Tries to build the program with the given options. See also the
1762C<$ctx->build> convenience function.
1763
1764If a callback is specified, then it will be called when compilation is
1765finished. Note that many OpenCL implementations block your program while
1766compiling whether you use a callback or not. See C<build_async> if you
1767want to make sure the build is done in the background.
1768
1769Note that some OpenCL implementations act up badly, and don't call the
1770callback in some error cases (but call it in others). This implementation
1771assumes the callback will always be called, and leaks memory if this is
1772not so. So best make sure you don't pass in invalid values.
1773
1774Some implementations fail with C<OpenCL::INVALID_BINARY> when the
1775compilation state is successful but some later stage fails.
1776
1777options: C<-D name>, C<-D name=definition>, C<-I dir>,
1778C<-cl-single-precision-constant>, C<-cl-denorms-are-zero>,
1779C<-cl-fp32-correctly-rounded-divide-sqrt>, C<-cl-opt-disable>,
1780C<-cl-mad-enable>, C<-cl-no-signed-zeros>, C<-cl-unsafe-math-optimizations>,
1781C<-cl-finite-math-only>, C<-cl-fast-relaxed-math>,
1782C<-w>, C<-Werror>, C<-cl-std=CL1.1/CL1.2>, C<-cl-kernel-arg-info>,
1783C<-create-library>, C<-enable-link-options>.
1784
1785build_status: OpenCL::BUILD_SUCCESS, OpenCL::BUILD_NONE,
1786OpenCL::BUILD_ERROR, OpenCL::BUILD_IN_PROGRESS.
626 1787
627L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clBuildProgram.html> 1788L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clBuildProgram.html>
1789
1790=item $program->build_async (\@devices = undef, $options = "", $cb->($program) = undef)
1791
1792Similar to C<< ->build >>, except it starts a thread, and never fails (you
1793need to check the compilation status form the callback, or by polling).
1794
1795=item $program->compile (\@devices = undef, $options = "", \%headers = undef, $cb->($program) = undef)
1796
1797Compiles the given program for the given devices (or all devices if
1798undef). If C<$headers> is given, it must be a hashref with include name =>
1799OpenCL::Program pairs.
1800
1801L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clCompileProgram.html>
628 1802
629=item $packed_value = $program->build_info ($device, $name) 1803=item $packed_value = $program->build_info ($device, $name)
630 1804
631Similar to C<< $platform->info >>, but returns build info for a previous 1805Similar to C<< $platform->info >>, but returns build info for a previous
632build attempt for the given device. 1806build attempt for the given device.
… …
638Creates an OpenCL::Kernel object out of the named C<__kernel> function in 1812Creates an OpenCL::Kernel object out of the named C<__kernel> function in
639the program. 1813the program.
640 1814
641L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateKernel.html> 1815L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateKernel.html>
642 1816
1817=item @kernels = $program->kernels_in_program
1818
1819Returns all kernels successfully compiled for all devices in program.
1820
1821http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clCreateKernelsInProgram.html
1822
1823=for gengetinfo begin program_build
1824
1825=item $build_status = $program->build_status ($device)
1826
1827Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_STATUS> and returns the result.
1828
1829=item $string = $program->build_options ($device)
1830
1831Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_OPTIONS> and returns the result.
1832
1833=item $string = $program->build_log ($device)
1834
1835Calls C<clGetProgramBuildInfo> with C<OpenCL::PROGRAM_BUILD_LOG> and returns the result.
1836
1837=for gengetinfo end program_build
1838
1839=item $packed_value = $program->info ($name)
1840
1841See C<< $platform->info >> for details.
1842
1843L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
1844
1845=for gengetinfo begin program
1846
1847=item $uint = $program->reference_count
1848
1849Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_REFERENCE_COUNT> and returns the result.
1850
1851=item $ctx = $program->context
1852
1853Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_CONTEXT> and returns the result.
1854
1855=item $uint = $program->num_devices
1856
1857Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_NUM_DEVICES> and returns the result.
1858
1859=item @devices = $program->devices
1860
1861Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_DEVICES> and returns the result.
1862
1863=item $string = $program->source
1864
1865Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_SOURCE> and returns the result.
1866
1867=item @ints = $program->binary_sizes
1868
1869Calls C<clGetProgramInfo> with C<OpenCL::PROGRAM_BINARY_SIZES> and returns the result.
1870
1871=for gengetinfo end program
1872
1873=item @blobs = $program->binaries
1874
1875Returns a string for the compiled binary for every device associated with
1876the program, empty strings indicate missing programs, and an empty result
1877means no program binaries are available.
1878
1879These "binaries" are often, in fact, informative low-level assembly
1880sources.
1881
1882L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProgramInfo.html>
1883
643=back 1884=back
644 1885
645=head2 THE OpenCL::Kernel CLASS 1886=head2 THE OpenCL::Kernel CLASS
646 1887
647=over 4 1888=over 4
… …
650 1891
651See C<< $platform->info >> for details. 1892See C<< $platform->info >> for details.
652 1893
653L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelInfo.html> 1894L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelInfo.html>
654 1895
1896=for gengetinfo begin kernel
1897
1898=item $string = $kernel->function_name
1899
1900Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_FUNCTION_NAME> and returns the result.
1901
1902=item $uint = $kernel->num_args
1903
1904Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_NUM_ARGS> and returns the result.
1905
1906=item $uint = $kernel->reference_count
1907
1908Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_REFERENCE_COUNT> and returns the result.
1909
1910=item $ctx = $kernel->context
1911
1912Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_CONTEXT> and returns the result.
1913
1914=item $program = $kernel->program
1915
1916Calls C<clGetKernelInfo> with C<OpenCL::KERNEL_PROGRAM> and returns the result.
1917
1918=for gengetinfo end kernel
1919
655=item $packed_value = $kernel->work_group_info ($device, $name) 1920=item $packed_value = $kernel->work_group_info ($device, $name)
1921
1922See C<< $platform->info >> for details.
1923
1924L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetKernelWorkGroupInfo.html>
1925
1926=for gengetinfo begin kernel_work_group
1927
1928=item $int = $kernel->work_group_size ($device)
1929
1930Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_WORK_GROUP_SIZE> and returns the result.
1931
1932=item @ints = $kernel->compile_work_group_size ($device)
1933
1934Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_COMPILE_WORK_GROUP_SIZE> and returns the result.
1935
1936=item $ulong = $kernel->local_mem_size ($device)
1937
1938Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_LOCAL_MEM_SIZE> and returns the result.
1939
1940=item $int = $kernel->preferred_work_group_size_multiple ($device)
1941
1942Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE> and returns the result.
1943
1944=item $ulong = $kernel->private_mem_size ($device)
1945
1946Calls C<clGetKernelWorkGroupInfo> with C<OpenCL::KERNEL_PRIVATE_MEM_SIZE> and returns the result.
1947
1948=for gengetinfo end kernel_work_group
1949
1950=item $packed_value = $kernel->arg_info ($idx, $name)
1951
1952See C<< $platform->info >> for details.
1953
1954L<http://www.khronos.org/registry/cl/sdk/1.2/docs/man/xhtml/clGetKernelArgInfo.html>
1955
1956=for gengetinfo begin kernel_arg
1957
1958=item $kernel_arg_address_qualifier = $kernel->arg_address_qualifier ($idx)
1959
1960Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_ADDRESS_QUALIFIER> and returns the result.
1961
1962=item $kernel_arg_access_qualifier = $kernel->arg_access_qualifier ($idx)
1963
1964Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_ACCESS_QUALIFIER> and returns the result.
1965
1966=item $string = $kernel->arg_type_name ($idx)
1967
1968Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_TYPE_NAME> and returns the result.
1969
1970=item $kernel_arg_type_qualifier = $kernel->arg_type_qualifier ($idx)
1971
1972Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_TYPE_QUALIFIER> and returns the result.
1973
1974=item $string = $kernel->arg_name ($idx)
1975
1976Calls C<clGetKernelArgInfo> with C<OpenCL::KERNEL_ARG_NAME> and returns the result.
1977
1978=for gengetinfo end kernel_arg
1979
1980=item $kernel->setf ($format, ...)
1981
1982Sets the arguments of a kernel. Since OpenCL 1.1 doesn't have a generic
1983way to set arguments (and with OpenCL 1.2 it might be rather slow), you
1984need to specify a format argument, much as with C<printf>, to tell OpenCL
1985what type of argument it is.
1986
1987The format arguments are single letters:
1988
1989 c char
1990 C unsigned char
1991 s short
1992 S unsigned short
1993 i int
1994 I unsigned int
1995 l long
1996 L unsigned long
1997
1998 h half float (0..65535)
1999 f float
2000 d double
2001
2002 z local (octet size)
2003
2004 m memory object (buffer or image)
2005 a sampler
2006 e event
2007
2008Space characters in the format string are ignored.
2009
2010Example: set the arguments for a kernel that expects an int, two floats, a buffer and an image.
2011
2012 $kernel->setf ("i ff mm", 5, 0.5, 3, $buffer, $image);
2013
2014=item $kernel->set_TYPE ($index, $value)
2015
2016=item $kernel->set_char ($index, $value)
2017
2018=item $kernel->set_uchar ($index, $value)
2019
2020=item $kernel->set_short ($index, $value)
2021
2022=item $kernel->set_ushort ($index, $value)
2023
2024=item $kernel->set_int ($index, $value)
2025
2026=item $kernel->set_uint ($index, $value)
2027
2028=item $kernel->set_long ($index, $value)
2029
2030=item $kernel->set_ulong ($index, $value)
2031
2032=item $kernel->set_half ($index, $value)
2033
2034=item $kernel->set_float ($index, $value)
2035
2036=item $kernel->set_double ($index, $value)
2037
2038=item $kernel->set_memory ($index, $value)
2039
2040=item $kernel->set_buffer ($index, $value)
2041
2042=item $kernel->set_image ($index, $value)
2043
2044=item $kernel->set_sampler ($index, $value)
2045
2046=item $kernel->set_local ($index, $value)
2047
2048=item $kernel->set_event ($index, $value)
2049
2050This is a family of methods to set the kernel argument with the number
2051C<$index> to the give C<$value>.
2052
2053Chars and integers (including the half type) are specified as integers,
2054float and double as floating point values, memory/buffer/image must be
2055an object of that type or C<undef>, local-memory arguments are set by
2056specifying the size, and sampler and event must be objects of that type.
2057
2058Note that C<set_memory> works for all memory objects (all types of buffers
2059and images) - the main purpose of the more specific C<set_TYPE> functions
2060is type checking.
2061
2062Setting an argument for a kernel does NOT keep a reference to the object -
2063for example, if you set an argument to some image object, free the image,
2064and call the kernel, you will run into undefined behaviour.
2065
2066L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetKernelArg.html>
2067
2068=back
2069
2070=head2 THE OpenCL::Event CLASS
2071
2072This is the superclass for all event objects (including OpenCL::UserEvent
2073objects).
2074
2075=over 4
2076
2077=item $ev->wait
2078
2079Waits for the event to complete.
2080
2081L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clWaitForEvents.html>
2082
2083=item $ev->cb ($exec_callback_type, $callback->($event, $event_command_exec_status))
2084
2085Adds a callback to the callback stack for the given event type. There is
2086no way to remove a callback again.
2087
2088L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetEventCallback.html>
2089
2090=item $packed_value = $ev->info ($name)
2091
2092See C<< $platform->info >> for details.
2093
2094L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetEventInfo.html>
2095
2096=for gengetinfo begin event
2097
2098=item $queue = $event->command_queue
2099
2100Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_QUEUE> and returns the result.
2101
2102=item $command_type = $event->command_type
2103
2104Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_TYPE> and returns the result.
2105
2106=item $uint = $event->reference_count
2107
2108Calls C<clGetEventInfo> with C<OpenCL::EVENT_REFERENCE_COUNT> and returns the result.
2109
2110=item $uint = $event->command_execution_status
2111
2112Calls C<clGetEventInfo> with C<OpenCL::EVENT_COMMAND_EXECUTION_STATUS> and returns the result.
2113
2114=item $ctx = $event->context
2115
2116Calls C<clGetEventInfo> with C<OpenCL::EVENT_CONTEXT> and returns the result.
2117
2118=for gengetinfo end event
2119
2120=item $packed_value = $ev->profiling_info ($name)
656 2121
657See C<< $platform->info >> for details. 2122See C<< $platform->info >> for details.
658 2123
659The reason this method is not called C<info> is that there already is an 2124The reason this method is not called C<info> is that there already is an
660C<< ->info >> method. 2125C<< ->info >> method.
661 2126
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> 2127L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetProfilingInfo.html>
2128
2129=for gengetinfo begin profiling
2130
2131=item $ulong = $event->profiling_command_queued
2132
2133Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_QUEUED> and returns the result.
2134
2135=item $ulong = $event->profiling_command_submit
2136
2137Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_SUBMIT> and returns the result.
2138
2139=item $ulong = $event->profiling_command_start
2140
2141Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_START> and returns the result.
2142
2143=item $ulong = $event->profiling_command_end
2144
2145Calls C<clGetEventProfilingInfo> with C<OpenCL::PROFILING_COMMAND_END> and returns the result.
2146
2147=for gengetinfo end profiling
678 2148
679=back 2149=back
680 2150
681=head2 THE OpenCL::Event CLASS 2151=head2 THE OpenCL::UserEvent CLASS
682 2152
683This is the superclass for all event objects (including OpenCL::UserEvent 2153This is a subclass of OpenCL::Event.
684objects).
685 2154
686=over 4 2155=over 4
687 2156
688=item $packed_value = $ev->info ($name) 2157=item $ev->set_status ($execution_status)
689 2158
690See C<< $platform->info >> for details. 2159Sets the execution status of the user event. Can only be called once,
2160either with OpenCL::COMPLETE or a negative number as status.
691 2161
692L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clGetEventInfo.html> 2162execution_status: OpenCL::COMPLETE or a negative integer.
693 2163
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> 2164L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetUserEventStatus.html>
708 2165
709=back 2166=back
710 2167
711=head2 THE OpenCL::UserEvent CLASS 2168=head2 THE OpenCL::Mapped CLASS
712 2169
713This is a subclass of OpenCL::Event. 2170This class represents objects mapped into host memory. They are
2171represented by a blessed string scalar. The string data is the mapped
2172memory area, that is, if you read or write it, then the mapped object is
2173accessed directly.
2174
2175You must only ever use operations that modify the string in-place - for
2176example, a C<substr> that doesn't change the length, or maybe a regex that
2177doesn't change the length. Any other operation might cause the data to be
2178copied.
2179
2180When the object is destroyed it will enqueue an implicit unmap operation
2181on the queue that was used to create it.
2182
2183Keep in mind that you I<need> to unmap (or destroy) mapped objects before
2184OpenCL sees the changes, even if some implementations don't need this
2185sometimes.
2186
2187Example, replace the first two floats in the mapped buffer by 1 and 2.
2188
2189 my $mapped = $queue->map_buffer ($buf, ...
2190 $mapped->event->wait; # make sure it's there
2191
2192 # now replace first 8 bytes by new data, which is exactly 8 bytes long
2193 # we blindly assume device endianness to equal host endianness
2194 # (and of course, we assume iee 754 single precision floats :)
2195 substr $$mapped, 0, 8, pack "f*", 1, 2;
714 2196
715=over 4 2197=over 4
716 2198
717=item $ev->set_status ($execution_status) 2199=item $ev = $mapped->unmap ($wait_events...)
718 2200
719L<http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clSetUserEventStatus.html> 2201Unmaps the mapped memory object, using the queue originally used to create
2202it, quite similarly to C<< $queue->unmap ($mapped, ...) >>.
2203
2204=item $bool = $mapped->mapped
2205
2206Returns whether the object is still mapped - true before an C<unmap> is
2207enqueued, false afterwards.
2208
2209=item $ev = $mapped->event
2210
2211Return the event object associated with the mapped object. Initially, this
2212will be the event object created when mapping the object, and after an
2213unmap, this will be the event object that the unmap operation created.
2214
2215=item $mapped->wait
2216
2217Same as C<< $mapped->event->wait >> - makes sure no operations on this
2218mapped object are outstanding.
2219
2220=item $bytes = $mapped->size
2221
2222Returns the size of the mapped area, in bytes. Same as C<length $$mapped>.
2223
2224=item $ptr = $mapped->ptr
2225
2226Returns the raw memory address of the mapped area.
2227
2228=item $mapped->set ($offset, $data)
2229
2230Replaces the data at the given C<$offset> in the memory area by the new
2231C<$data>. This method is safer than direct manipulation of C<$mapped>
2232because it does bounds-checking, but also slower.
2233
2234=item $data = $mapped->get ($offset, $length)
2235
2236Returns (without copying) a scalar representing the data at the given
2237C<$offset> and C<$length> in the mapped memory area. This is the same as
2238the following substr, except much slower;
2239
2240 $data = substr $$mapped, $offset, $length
2241
2242=cut
2243
2244sub OpenCL::Mapped::get {
2245 substr ${$_[0]}, $_[1], $_[2]
2246}
720 2247
721=back 2248=back
722 2249
2250=head2 THE OpenCL::MappedBuffer CLASS
2251
2252This is a subclass of OpenCL::Mapped, representing mapped buffers.
2253
2254=head2 THE OpenCL::MappedImage CLASS
2255
2256This is a subclass of OpenCL::Mapped, representing mapped images.
2257
2258=over 4
2259
2260=item $bytes = $mapped->row_pitch
2261
2262=item $bytes = $mapped->slice_pitch
2263
2264Return the row or slice pitch of the image that has been mapped.
2265
2266=back
2267
2268
723=cut 2269=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 2270
7441; 22711;
745 2272
746=head1 AUTHOR 2273=head1 AUTHOR
747 2274

Diff Legend

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