[Beignet] [hpc12/tools] build of 'sum' on 'Intel(R) HD Graphics IvyBridge M GT2' failed

[Zhigang Gong]

Just got some time to try you example with PyOpenCL. And it seems that
PyOpenCL depends on
a function which we haven't implemented. After apply the following
patch, I can run your example
correctly. The output is as below:
gongzg at gongzg-MBA:~/Downloads$ python test.py
('a', array([ 0.], dtype=float32))
('b', array([ 0.], dtype=float32))
('c', array([ 0.], dtype=float32))
1 0.010514
('a', array([ 0.,  1.], dtype=float32))
('b', array([ 0.,  1.], dtype=float32))
('c', array([ 0.,  2.], dtype=float32))
2 0.004235
('a', array([ 0.,  1.,  2.], dtype=float32))
('b', array([ 0.,  1.,  2.], dtype=float32))
('c', array([ 0.,  2.,  4.], dtype=float32))
3 0.004166


>From 81c9e5cf70ec01197833f8722f6aa16632a5f1a4 Mon Sep 17 00:00:00 2001
From: Zhigang Gong <zhigang.gong at linux.intel.com>
Date: Sat, 30 Aug 2014 16:34:44 +0800
Subject: [PATCH] Runtime: Implement clGetExtensionFunctionAddressForPlatform.

It seems that this function is required for PyOpenCL.

Signed-off-by: Zhigang Gong <zhigang.gong at linux.intel.com>
---
 src/cl_api.c     | 19 +++++++++++++++++--
 src/cl_khr_icd.c |  2 +-
 2 files changed, 18 insertions(+), 3 deletions(-)

diff --git a/src/cl_api.c b/src/cl_api.c
index 177a7e8..b463128 100644
--- a/src/cl_api.c
+++ b/src/cl_api.c
@@ -3156,8 +3156,8 @@ error:
   if (strcmp(#x, func_name) == 0)       \
     return (void *)x;

-void*
-clGetExtensionFunctionAddress(const char *func_name)
+static void*
+internal_clGetExtensionFunctionAddress(const char *func_name)
 {
   if (func_name == NULL)
     return NULL;
@@ -3180,6 +3180,21 @@ clGetExtensionFunctionAddress(const char *func_name)
   return NULL;
 }

+void*
+clGetExtensionFunctionAddress(const char *func_name)
+{
+  return internal_clGetExtensionFunctionAddress(func_name);
+}
+
+void*
+clGetExtensionFunctionAddressForPlatform(cl_platform_id platform,
+                              const char *func_name)
+{
+  if (UNLIKELY(platform != NULL && platform != intel_platform))
+    return NULL;
+  return internal_clGetExtensionFunctionAddress(func_name);
+}
+
 #undef EXTFUNC

 cl_int
diff --git a/src/cl_khr_icd.c b/src/cl_khr_icd.c
index 6d49db0..50a0898 100644
--- a/src/cl_khr_icd.c
+++ b/src/cl_khr_icd.c
@@ -154,7 +154,7 @@ struct _cl_icd_dispatch const cl_khr_icd_dispatch = {
   clEnqueueMigrateMemObjects,
   clEnqueueMarkerWithWaitList,
   clEnqueueBarrierWithWaitList,
-  CL_1_2_NOTYET(clGetExtensionFunctionAddressForPlatform),
+  clGetExtensionFunctionAddressForPlatform,
   CL_GL_INTEROP(clCreateFromGLTexture),
   (void *) NULL,
   (void *) NULL,
-- 
1.9.1


On Sat, Aug 30, 2014 at 9:08 AM, David Liebman
<david.c.liebman at gmail.com> wrote:
>
> On Aug 29, 2014 8:26 PM, "Zhigang Gong" <zhigang.gong at gmail.com> wrote:
>>
>> I guess you may only change the kernel. Please also replace all the double
>> with float in the cl-demo.c.Then rebuild and run it again. I just tried, and
>> it works well.
>>
>>
>> _______________________________________________
>> Beignet mailing list
>> Beignet at lists.freedesktop.org
>> http://lists.freedesktop.org/mailman/listinfo/beignet
>>
> How about this program? It has no double. C always prints zero.  It's
> pyopencl.
>
> import pyopencl as cl
> import numpy
> import sys
> import time
>
> class CL(object):
>     def __init__(self, size=10):
>         self.size = size
>         self.ctx = cl.create_some_context()
>         self.queue = cl.CommandQueue(self.ctx)
>
>     def load_program(self):
>         fstr="""
>         __kernel void part1(__global float* a, __global float* b, __global
> float* c)
>         {
>             unsigned int i = get_global_id(0);
>
>            c[i] = a[i] + b[i];
>         }
>          """
>         self.program = cl.Program(self.ctx, fstr).build()
>
>     def popCorn(self):
>         mf = cl.mem_flags
>
>         self.a = numpy.array(range(self.size), dtype=numpy.float32)
>         self.b = numpy.array(range(self.size), dtype=numpy.float32)
>
>         self.a_buf = cl.Buffer(self.ctx, mf.READ_ONLY | mf.COPY_HOST_PTR,
>                                hostbuf=self.a)
>         self.b_buf = cl.Buffer(self.ctx, mf.READ_ONLY | mf.COPY_HOST_PTR,
>                                hostbuf=self.b)
>         self.dest_buf = cl.Buffer(self.ctx, mf.WRITE_ONLY, self.b.nbytes)
>
>     def execute(self):
>         self.program.part1(self.queue, self.a.shape, None, self.a_buf,
> self.b_buf, self.dest_buf)
>         c = numpy.empty_like(self.a)
>         cl.enqueue_read_buffer(self.queue, self.dest_buf, c).wait()
>         print ( "a", self.a)
>         print ( "b", self.b)
>         print ( "c", c )
>
> def add(s=10) :
>     starttime = time.clock()
>     matrixmul = CL(s)
>     matrixmul.load_program()
>     matrixmul.popCorn()
>     matrixmul.execute()
>     endtime = time.clock()
>     print s, endtime - starttime
>
> if __name__ == '__main__':
>         #add(1)
>         #add(2)
>         #add(3)
>         #add(4)
>         #add(5)
>         add(50)
>         add(500)
>         add(5000)
>         add(50000)
>         add(500000)
>
>         #add(10000000)
>