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

Comparing OpenCL/OpenCL.pm (file contents):
Revision 1.1 by root, Tue Nov 15 06:50:30 2011 UTC vs.
Revision 1.71 by root, Fri May 4 14:29:35 2012 UTC

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

Diff Legend

Removed lines
+ Added lines
< Changed lines
> Changed lines