Skip to content
Browse files

Implement cl_create_program_with_source().

  • Loading branch information...
1 parent 08d6e94 commit 2cbb1355745c715460cf1290f17e2ed50720a8fe @rsky committed Sep 1, 2012
Showing with 112 additions and 15 deletions.
  1. +38 −0 examples/mandelbrot.cl
  2. +4 −0 examples/mandelbrot.php
  3. +22 −15 opencl.c
  4. +47 −0 program.c
  5. +1 −0 program.h
View
38 examples/mandelbrot.cl
@@ -0,0 +1,38 @@
+__kernel
+void Mandelbrot(
+ __global unsigned char *output,
+ const int w,
+ const int h,
+ const float cx,
+ const float cy,
+ const float unit)
+{
+ int globalID = get_global_id(0);
+ int ox = globalID % w;
+ int oy = globalID / w;
+ int ix = ox;
+ int iy = h - 1 - oy;
+
+ if ( ix >= w || iy >= h ) { return; }
+ float fx = (float)(ix - w / 2) * unit + cx;
+ float fy = (float)(iy - h / 2) * unit + cy;
+
+ float r = fx;
+ float i = fy;
+ int n;
+ int m = 200;
+ for (n = 0; n < m; n++) {
+ float rr = r * r;
+ float ii = i * i;
+ float ri = r * i;
+ r = fx + rr - ii;
+ i = fy + 2 * ri;
+ if ( rr + ii > 4 ) { break; }
+ }
+ float fval = (float)n / (float)m;
+ int ival = 256 * fval;
+ if (ival < 0) { ival = 0; }
+ if (ival > 255) { ival = 255; }
+
+ output[ox + oy * w] = (unsigned char)ival;
+}
View
4 examples/mandelbrot.php
@@ -10,3 +10,7 @@
$queue = cl_create_command_queue($context);
var_dump($queue);
+
+$source = __DIR__ . '/mandelbrot.cl';
+$program = cl_create_program_with_source($context, $source, $err);
+var_dump($program, $err);
View
37 opencl.c
@@ -114,6 +114,12 @@ ZEND_BEGIN_ARG_INFO_EX(arginfo_get_program_info, ZEND_SEND_BY_VAL, ZEND_RETURN_V
ZEND_ARG_INFO(0, name)
ZEND_END_ARG_INFO()
+ZEND_BEGIN_ARG_INFO_EX(arginfo_create_program_with_source, ZEND_SEND_BY_VAL, ZEND_RETURN_VALUE, 2)
+ ZEND_ARG_INFO(0, context)
+ ZEND_ARG_INFO(0, string)
+ ZEND_ARG_INFO(1, errcode)
+ZEND_END_ARG_INFO()
+
/* kernel */
ZEND_BEGIN_ARG_INFO_EX(arginfo_get_kernel_info, ZEND_SEND_BY_VAL, ZEND_RETURN_VALUE, 1)
ZEND_ARG_INFO(0, kernel)
@@ -137,28 +143,29 @@ ZEND_END_ARG_INFO()
static zend_function_entry phpcl_functions[] = {
/* platform */
- PHP_FE(cl_get_platform_info, arginfo_get_platform_info)
- PHP_FE(cl_get_platform_ids, NULL)
+ PHP_FE(cl_get_platform_info, arginfo_get_platform_info)
+ PHP_FE(cl_get_platform_ids, NULL)
/* device */
- PHP_FE(cl_get_device_info, arginfo_get_device_info)
- PHP_FE(cl_get_device_ids, arginfo_get_device_ids)
+ PHP_FE(cl_get_device_info, arginfo_get_device_info)
+ PHP_FE(cl_get_device_ids, arginfo_get_device_ids)
/* context */
- PHP_FE(cl_get_context_info, arginfo_get_context_info)
- PHP_FE(cl_create_context, arginfo_create_context)
+ PHP_FE(cl_get_context_info, arginfo_get_context_info)
+ PHP_FE(cl_create_context, arginfo_create_context)
/* command queue */
- PHP_FE(cl_get_command_queue_info, arginfo_get_command_queue_info)
- PHP_FE(cl_create_command_queue, arginfo_create_command_queue)
+ PHP_FE(cl_get_command_queue_info, arginfo_get_command_queue_info)
+ PHP_FE(cl_create_command_queue, arginfo_create_command_queue)
/* mem */
- PHP_FE(cl_get_mem_object_info, arginfo_get_mem_object_info)
- /*PHP_FE(cl_get_image_info, arginfo_get_image_info)*/
+ PHP_FE(cl_get_mem_object_info, arginfo_get_mem_object_info)
+ /*PHP_FE(cl_get_image_info, arginfo_get_image_info)*/
/* program */
- PHP_FE(cl_get_program_info, arginfo_get_program_info)
+ PHP_FE(cl_get_program_info, arginfo_get_program_info)
+ PHP_FE(cl_create_program_with_source, arginfo_create_program_with_source)
/* kernel */
- PHP_FE(cl_get_kernel_info, arginfo_get_kernel_info)
+ PHP_FE(cl_get_kernel_info, arginfo_get_kernel_info)
/* event */
- PHP_FE(cl_get_event_info, arginfo_get_event_info)
+ PHP_FE(cl_get_event_info, arginfo_get_event_info)
/* sampler */
- PHP_FE(cl_get_sampler_info, arginfo_get_sampler_info)
+ PHP_FE(cl_get_sampler_info, arginfo_get_sampler_info)
/* terminate */
{ NULL, NULL, NULL }
};
@@ -355,7 +362,7 @@ int phpcl_le_event(void)
int phpcl_le_program(void)
{
- return le_platform;
+ return le_program;
}
int phpcl_le_kernel(void)
View
47 program.c
@@ -105,6 +105,53 @@ PHP_FUNCTION(cl_get_program_info)
}
/* }}} */
+/* {{{ resource cl_program cl_create_program_with_source(resource cl_context context, mixed source[, int &errcode]); */
+
+PHP_FUNCTION(cl_create_program_with_source)
+{
+ cl_int errcode = CL_SUCCESS;
+ cl_program program = NULL;
+ zval *zcontext = NULL;
+ phpcl_context_t *ctx = NULL;
+ zval *zstring = NULL;
+ zval *zerrcode = NULL;
+
+ RETVAL_FALSE;
+
+ if (zend_parse_parameters(ZEND_NUM_ARGS() TSRMLS_CC,
+ "rz|z", &zcontext, &zstring, &zerrcode) == FAILURE) {
+ return;
+ }
+
+ ZEND_FETCH_RESOURCE(ctx, phpcl_context_t *, &zcontext, -1,
+ "cl_context", phpcl_le_context());
+
+ if (Z_TYPE_P(zstring) == IS_STRING) {
+ const char *strings[] = { (const char *)Z_STRVAL_P(zstring) };
+ const size_t lengths[] = { (const size_t)Z_STRLEN_P(zstring) };
+ program = clCreateProgramWithSource(ctx->context, 1, strings, lengths, &errcode);
+ } else if (Z_TYPE_P(zstring) == IS_ARRAY) {
+ /* TODO: support multiple sources
+ cl_uint count = 0;
+ char **strings = NULL;
+ size_t *lengths = NULL;
+ program = clCreateProgramWithSource(ctx->context, count, strings, lengths, &errcode);
+ efree(strings);
+ efree(lengths);
+ */
+ }
+
+ if (zerrcode) {
+ zval_dtor(zerrcode);
+ ZVAL_LONG(zerrcode, (long)errcode);
+ }
+
+ if (program) {
+ ZEND_REGISTER_RESOURCE(return_value, program, phpcl_le_program());
+ }
+}
+
+/* }}} */
/*
* Local variables:
View
1 program.h
@@ -13,6 +13,7 @@
#define PHPCL_PROGRAM_H
PHP_FUNCTION(cl_get_program_info);
+PHP_FUNCTION(cl_create_program_with_source);
#endif

0 comments on commit 2cbb135

Please sign in to comment.
Something went wrong with that request. Please try again.