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

Comparing OpenCL/OpenCL.xs (file contents):
Revision 1.3 by root, Tue Nov 15 20:38:07 2011 UTC vs.
Revision 1.6 by root, Thu Nov 17 01:36:52 2011 UTC

… …
20typedef cl_mem OpenCL__Image3D_ornull; 20typedef cl_mem OpenCL__Image3D_ornull;
21typedef cl_sampler OpenCL__Sampler; 21typedef cl_sampler OpenCL__Sampler;
22typedef cl_program OpenCL__Program; 22typedef cl_program OpenCL__Program;
23typedef cl_kernel OpenCL__Kernel; 23typedef cl_kernel OpenCL__Kernel;
24typedef cl_event OpenCL__Event; 24typedef cl_event OpenCL__Event;
25typedef cl_event OpenCL__UserEvent;
26
27/*****************************************************************************/
28
29/* up to two temporary buffers */
30static void *
31tmpbuf (size_t size)
32{
33 static int idx;
34 static void *buf [2];
35 static size_t len [2];
36
37 idx ^= 1;
38
39 if (len [idx] < size)
40 {
41 free (buf [idx]);
42 len [idx] = ((size + 31) & ~4095) + 4096 - 32;
43 buf [idx] = malloc (len [idx]);
44 }
45
46 return buf [idx];
47}
48
49/*****************************************************************************/
25 50
26typedef struct 51typedef struct
27{ 52{
28 IV iv; 53 IV iv;
29 const char *name; 54 const char *name;
… …
63 }; 88 };
64 89
65 return iv2str (err, errstr, sizeof (errstr) / sizeof (errstr [0]), "ERROR(%d)"); 90 return iv2str (err, errstr, sizeof (errstr) / sizeof (errstr [0]), "ERROR(%d)");
66} 91}
67 92
93/*****************************************************************************/
94
95static cl_int last_error;
96
68#define FAIL(name,res) \ 97#define FAIL(name,err) \
69 croak ("cl" # name ": %s", err2str (res)); 98 croak ("cl" # name ": %s", err2str (last_error = err));
70 99
71#define NEED_SUCCESS(name,args) \ 100#define NEED_SUCCESS(name,args) \
72 do { \ 101 do { \
73 cl_int res = cl ## name args; \ 102 cl_int res = cl ## name args; \
74 \ 103 \
75 if (res) \ 104 if (res) \
76 FAIL (name, res); \ 105 FAIL (name, res); \
77 } while (0) 106 } while (0)
78 107
108/*****************************************************************************/
109
79#define NEW_MORTAL_OBJ(class,ptr) sv_setref_pv (sv_newmortal (), class, ptr) 110#define NEW_MORTAL_OBJ(class,ptr) sv_setref_pv (sv_newmortal (), class, ptr)
80#define XPUSH_NEW_OBJ(class,ptr) XPUSHs (NEW_MORTAL_OBJ (class, ptr)) 111#define XPUSH_NEW_OBJ(class,ptr) XPUSHs (NEW_MORTAL_OBJ (class, ptr))
81 112
82/*TODO*/ 113static void *
83#define EVENT_LIST(items,count) cl_uint event_list_count = 0; cl_event *event_list_ptr = 0 114SvPTROBJ (const char *func, const char *svname, SV *sv, const char *pkg)
115{
116 if (SvROK (sv) && sv_derived_from (sv, pkg))
117 return (void *)SvIV (SvRV (sv));
118
119 croak ("%s: %s is not of type %s", func, svname, pkg);
120}
121
122/*****************************************************************************/
123
124static cl_event *
125event_list (SV **items, int count)
126{
127 cl_event *list = tmpbuf (sizeof (cl_event) * count);
128
129 while (count--)
130 list [count] = SvPTROBJ ("clEnqueue", "wait_events", items [count], "OpenCL::Event");
131
132 return list;
133}
134
135#define EVENT_LIST(items,count) \
136 cl_uint event_list_count = (count); \
137 cl_event *event_list_ptr = event_list (&ST (items), event_list_count)
84 138
85#define INFO(class) \ 139#define INFO(class) \
86{ \ 140{ \
87 size_t size; \ 141 size_t size; \
88 SV *sv; \ 142 SV *sv; \
… …
94 SvCUR_set (sv, size); \ 148 SvCUR_set (sv, size); \
95 NEED_SUCCESS (Get ## class ## Info, (this, name, size, SvPVX (sv), 0)); \ 149 NEED_SUCCESS (Get ## class ## Info, (this, name, size, SvPVX (sv), 0)); \
96 XPUSHs (sv); \ 150 XPUSHs (sv); \
97} 151}
98 152
99static void *
100SvPTROBJ (const char *func, const char *svname, SV *sv, const char *pkg)
101{
102 if (SvROK (sv) && sv_derived_from (sv, pkg))
103 return (void *)SvIV (SvRV (sv));
104
105 croak ("%s: %s is not of type %s", func, svname, pkg);
106}
107
108MODULE = OpenCL PACKAGE = OpenCL 153MODULE = OpenCL PACKAGE = OpenCL
109 154
110PROTOTYPES: ENABLE 155PROTOTYPES: ENABLE
111 156
112BOOT: 157BOOT:
113{ 158{
114 HV *stash = gv_stashpv ("OpenCL", 1); 159 HV *stash = gv_stashpv ("OpenCL", 1);
115 static const ivstr *civ, const_iv[] = { 160 static const ivstr *civ, const_iv[] = {
116 { sizeof (cl_char ), "SIZEOF_CHAR" }, 161 { sizeof (cl_char ), "SIZEOF_CHAR" },
117 { sizeof (cl_uchar ), "SIZEOF_UCHAR" }, 162 { sizeof (cl_uchar ), "SIZEOF_UCHAR" },
118 { sizeof (cl_short ), "SIZEOF_SHORT" }, 163 { sizeof (cl_short ), "SIZEOF_SHORT" },
119 { sizeof (cl_ushort), "SIZEOF_USHORT"}, 164 { sizeof (cl_ushort), "SIZEOF_USHORT" },
120 { sizeof (cl_int ), "SIZEOF_INT" }, 165 { sizeof (cl_int ), "SIZEOF_INT" },
121 { sizeof (cl_uint ), "SIZEOF_UINT" }, 166 { sizeof (cl_uint ), "SIZEOF_UINT" },
122 { sizeof (cl_long ), "SIZEOF_LONG" }, 167 { sizeof (cl_long ), "SIZEOF_LONG" },
123 { sizeof (cl_ulong ), "SIZEOF_ULONG" }, 168 { sizeof (cl_ulong ), "SIZEOF_ULONG" },
124 { sizeof (cl_half ), "SIZEOF_HALF" }, 169 { sizeof (cl_half ), "SIZEOF_HALF" },
125 { sizeof (cl_float ), "SIZEOF_FLOAT" }, 170 { sizeof (cl_float ), "SIZEOF_FLOAT" },
171 { sizeof (cl_double), "SIZEOF_DOUBLE" },
126#include "constiv.h" 172#include "constiv.h"
127 }; 173 };
128 for (civ = const_iv + sizeof (const_iv) / sizeof (const_iv [0]); civ > const_iv; civ--) 174 for (civ = const_iv + sizeof (const_iv) / sizeof (const_iv [0]); civ > const_iv; civ--)
129 newCONSTSUB (stash, (char *)civ[-1].name, newSViv (civ[-1].iv)); 175 newCONSTSUB (stash, (char *)civ[-1].name, newSViv (civ[-1].iv));
130} 176}
131 177
178cl_int
179errno ()
180 CODE:
181 errno = last_error;
182
132const char * 183const char *
133err2str (cl_int err) 184err2str (cl_int err)
134 185
135const char * 186const char *
136enum2str (cl_uint value) 187enum2str (cl_uint value)
… …
142 cl_platform_id *list; 193 cl_platform_id *list;
143 cl_uint count; 194 cl_uint count;
144 int i; 195 int i;
145 196
146 NEED_SUCCESS (GetPlatformIDs, (0, 0, &count)); 197 NEED_SUCCESS (GetPlatformIDs, (0, 0, &count));
147 Newx (list, count, cl_platform_id); 198 list = tmpbuf (sizeof (*list) * count);
148 NEED_SUCCESS (GetPlatformIDs, (count, list, 0)); 199 NEED_SUCCESS (GetPlatformIDs, (count, list, 0));
149 200
150 EXTEND (SP, count); 201 EXTEND (SP, count);
151 for (i = 0; i < count; ++i) 202 for (i = 0; i < count; ++i)
152 PUSHs (NEW_MORTAL_OBJ ("OpenCL::Platform", list [i])); 203 PUSHs (NEW_MORTAL_OBJ ("OpenCL::Platform", list [i]));
153
154 Safefree (list);
155} 204}
156 205
157void 206void
158context_from_type_simple (cl_device_type type = CL_DEVICE_TYPE_DEFAULT) 207context_from_type_simple (cl_device_type type = CL_DEVICE_TYPE_DEFAULT)
159 PPCODE: 208 PPCODE:
… …
191 cl_device_id *list; 240 cl_device_id *list;
192 cl_uint count; 241 cl_uint count;
193 int i; 242 int i;
194 243
195 NEED_SUCCESS (GetDeviceIDs, (this, type, 0, 0, &count)); 244 NEED_SUCCESS (GetDeviceIDs, (this, type, 0, 0, &count));
196 Newx (list, count, cl_device_id); 245 list = tmpbuf (sizeof (*list) * count);
197 NEED_SUCCESS (GetDeviceIDs, (this, type, count, list, 0)); 246 NEED_SUCCESS (GetDeviceIDs, (this, type, count, list, 0));
198 247
199 EXTEND (SP, count); 248 EXTEND (SP, count);
200 for (i = 0; i < count; ++i) 249 for (i = 0; i < count; ++i)
201 PUSHs (sv_setref_pv (sv_newmortal (), "OpenCL::Device", list [i])); 250 PUSHs (sv_setref_pv (sv_newmortal (), "OpenCL::Device", list [i]));
202
203 Safefree (list);
204} 251}
205 252
206void 253void
207context_from_type_simple (OpenCL::Platform this, cl_device_type type = CL_DEVICE_TYPE_DEFAULT) 254context_from_type_simple (OpenCL::Platform this, cl_device_type type = CL_DEVICE_TYPE_DEFAULT)
208 PPCODE: 255 PPCODE:
… …
248info (OpenCL::Context this, cl_context_info name) 295info (OpenCL::Context this, cl_context_info name)
249 PPCODE: 296 PPCODE:
250 INFO (Context) 297 INFO (Context)
251 298
252void 299void
253command_queue_simple (OpenCL::Context this, OpenCL::Device device) 300queue (OpenCL::Context this, OpenCL::Device device, cl_command_queue_properties properties)
254 PPCODE: 301 PPCODE:
255{ 302{
256 cl_int res; 303 cl_int res;
257 cl_command_queue queue = clCreateCommandQueue (this, device, 0, &res); 304 cl_command_queue queue = clCreateCommandQueue (this, device, properties, &res);
258 305
259 if (res) 306 if (res)
260 FAIL (CreateCommandQueue, res); 307 FAIL (CreateCommandQueue, res);
261 308
262 XPUSH_NEW_OBJ ("OpenCL::Queue", queue); 309 XPUSH_NEW_OBJ ("OpenCL::Queue", queue);
310}
311
312void
313user_event (OpenCL::Context this)
314 PPCODE:
315{
316 cl_int res;
317 cl_event ev = clCreateUserEvent (this, &res);
318
319 if (res)
320 FAIL (CreateUserevent, res);
321
322 XPUSH_NEW_OBJ ("OpenCL::UserEvent", ev);
263} 323}
264 324
265void 325void
266buffer (OpenCL::Context this, cl_mem_flags flags, size_t len) 326buffer (OpenCL::Context this, cl_mem_flags flags, size_t len)
267 PPCODE: 327 PPCODE:
… …
558 618
559 if (ev) 619 if (ev)
560 XPUSH_NEW_OBJ ("OpenCL::Event", ev); 620 XPUSH_NEW_OBJ ("OpenCL::Event", ev);
561} 621}
562 622
563 /*TODO http://www.khronos.org/registry/cl/sdk/1.1/docs/man/xhtml/clEnqueueNDRangeKernel.html */ 623void
624enqueue_nd_range_kernel (OpenCL::Queue this, OpenCL::Kernel kernel, SV *global_work_offset, SV *global_work_size, SV *local_work_size = &PL_sv_undef, ...)
625 PPCODE:
626{
627 cl_event ev = 0;
628 size_t *gwo = 0, *gws, *lws = 0;
629 int gws_len;
630 size_t *lists;
631 int i;
632 EVENT_LIST (5, items - 5);
633
634 if (!SvROK (global_work_size) || SvTYPE (SvRV (global_work_size)) != SVt_PVAV)
635 croak ("clEnqueueNDRangeKernel: global_work_size must be an array reference");
636
637 gws_len = AvFILLp (SvRV (global_work_size)) + 1;
638
639 lists = tmpbuf (sizeof (size_t) * 3 * gws_len);
640
641 gws = lists + gws_len * 0;
642 for (i = 0; i < gws_len; ++i)
643 gws [i] = SvIV (AvARRAY (SvRV (global_work_size))[i]);
644
645 if (SvOK (global_work_offset))
646 {
647 if (!SvROK (global_work_offset) || SvTYPE (SvRV (global_work_offset)) != SVt_PVAV)
648 croak ("clEnqueueNDRangeKernel: global_work_offset must be undef or an array reference");
649
650 if (AvFILLp (SvRV (global_work_size)) + 1 != gws_len)
651 croak ("clEnqueueNDRangeKernel: global_work_offset must be undef or an array of same size as global_work_size");
652
653 gwo = lists + gws_len * 1;
654 for (i = 0; i < gws_len; ++i)
655 gwo [i] = SvIV (AvARRAY (SvRV (global_work_offset))[i]);
656 }
657
658 if (SvOK (local_work_size))
659 {
660 if (SvOK (local_work_size) && !SvROK (local_work_size) || SvTYPE (SvRV (local_work_size)) != SVt_PVAV)
661 croak ("clEnqueueNDRangeKernel: global_work_size must be undef or an array reference");
662
663 if (AvFILLp (SvRV (local_work_size)) + 1 != gws_len)
664 croak ("clEnqueueNDRangeKernel: local_work_local must be undef or an array of same size as global_work_size");
665
666 lws = lists + gws_len * 2;
667 for (i = 0; i < gws_len; ++i)
668 lws [i] = SvIV (AvARRAY (SvRV (local_work_size))[i]);
669 }
670
671 NEED_SUCCESS (EnqueueNDRangeKernel, (this, kernel, gws_len, gwo, gws, lws, event_list_count, event_list_ptr, GIMME_V != G_VOID ? &ev : 0));
672
673 if (ev)
674 XPUSH_NEW_OBJ ("OpenCL::Event", ev);
675}
564 676
565void 677void
566enqueue_marker (OpenCL::Queue this) 678enqueue_marker (OpenCL::Queue this)
567 PPCODE: 679 PPCODE:
568{ 680{
… …
725set_float (OpenCL::Kernel this, cl_uint idx, cl_float value) 837set_float (OpenCL::Kernel this, cl_uint idx, cl_float value)
726 CODE: 838 CODE:
727 clSetKernelArg (this, idx, sizeof (value), &value); 839 clSetKernelArg (this, idx, sizeof (value), &value);
728 840
729void 841void
842set_double (OpenCL::Kernel this, cl_uint idx, cl_double value)
843 CODE:
844 clSetKernelArg (this, idx, sizeof (value), &value);
845
846void
730set_memory (OpenCL::Kernel this, cl_uint idx, OpenCL::Memory_ornull value) 847set_memory (OpenCL::Kernel this, cl_uint idx, OpenCL::Memory_ornull value)
731 CODE: 848 CODE:
732 clSetKernelArg (this, idx, sizeof (value), &value); 849 clSetKernelArg (this, idx, sizeof (value), &value);
733 850
734void 851void
… …
771void 888void
772wait (OpenCL::Event this) 889wait (OpenCL::Event this)
773 CODE: 890 CODE:
774 clWaitForEvents (1, &this); 891 clWaitForEvents (1, &this);
775 892
893MODULE = OpenCL PACKAGE = OpenCL::UserEvent
894
895void
896set_status (OpenCL::UserEvent this, cl_int execution_status)
897 CODE:
898 clSetUserEventStatus (this, execution_status);
899

Diff Legend

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