ViewVC Help
View File | Revision Log | Show Annotations | Download File
/cvs/OpenCL/OpenCL.pm
Revision: 1.74
Committed: Fri May 4 15:04:51 2012 UTC (12 years ago) by root
Branch: MAIN
Changes since 1.73: +8 -1 lines
Log Message:
*** empty log message ***

File Contents

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