initial import of CLProgram. Updated JUnit Test accordingly.

6 files changed, 358 insertions, 64 deletions

diff --git a/src/com/mbien/opencl/CLContext.java b/src/com/mbien/opencl/CLContext.java
index 203172b0..601a2b79 100644
--- a/src/com/mbien/opencl/CLContext.java
+++ b/src/com/mbien/opencl/CLContext.java
@@ -1,18 +1,28 @@
 package com.mbien.opencl;
 
 import com.mbien.opencl.impl.CLImpl;
-import com.sun.gluegen.runtime.PointerBuffer;
+import java.nio.ByteBuffer;
+import java.nio.ByteOrder;
 import java.nio.IntBuffer;
 
+import java.util.ArrayList;
+import java.util.Collections;
+import java.util.List;
+import static com.mbien.opencl.CLException.*;
+
 /**
  *
  * @author Michael Bien
  */
 public final class CLContext {
 
-    private final static CL cl;
+    final static CL cl;
     public final long contextID;
 
+    private CLDevice[] devices;
+
+    private final List<CLProgram> programs;
+
     static{
         System.loadLibrary("gluegen-rt");
         System.loadLibrary("jocl");
@@ -21,45 +31,66 @@ public final class CLContext {
 
     private CLContext(long contextID) {
         this.contextID = contextID;
+        this.programs = new ArrayList<CLProgram>();
     }
 
     /**
      * Creates a default context on all available devices.
      */
-    public static CLContext create() {
-        IntBuffer ib = IntBuffer.allocate(1);
-        long context = cl.clCreateContextFromType(null, 0, CL.CL_DEVICE_TYPE_ALL, null, null, null, 0);
-
-//        int errorCode = ib.get();
-//        if(errorCode != CL.CL_SUCCESS)
-//            throw new CLException(errorCode, "can not create CL context");
-
-        return new CLContext(context);
+    public static final CLContext create() {
+        return createContext(CL.CL_DEVICE_TYPE_ALL);
     }
 
     /**
      * Creates a default context on the specified device types.
      */
-    public static CLContext create(CLDevice.Type... deviceTypes) {
+    public static final CLContext create(CLDevice.Type... deviceTypes) {
 
         int type = deviceTypes[0].CL_TYPE;
         for (int i = 1; i < deviceTypes.length; i++) {
             type |= deviceTypes[i].CL_TYPE;
         }
 
-        long ctxID = cl.clCreateContextFromType(null, 0, type, null, null, null, 0);
-        return new CLContext(ctxID);
+        return createContext(type);
+    }
+
+    private static final CLContext createContext(long deviceType) {
+
+        IntBuffer error = IntBuffer.allocate(1);
+        long context = cl.clCreateContextFromType(null, 0, deviceType, null, null, error, 0);
+
+        checkForError(error.get(), "can not create CL context");
+
+        return new CLContext(context);
+    }
+    
+    public CLProgram createProgram(String src) {
+        CLProgram program = new CLProgram(this, src, contextID);
+        programs.add(program);
+        return program;
+    }
+
+    void programReleased(CLProgram program) {
+        programs.remove(program);
     }
 
     /**
      * Releases the context and all resources.
      */
-    public void release() {
+    public CLContext release() {
         int ret = cl.clReleaseContext(contextID);
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "error releasing context");
+        checkForError(ret, "error releasing context");
+        return this;
+    }
+
+    /**
+     * Returns a read only view of all programs associated with this context.
+     */
+    public List<CLProgram> getPrograms() {
+        return Collections.unmodifiableList(programs);
     }
 
+
     /**
      * Gets the device with maximal FLOPS from this context.
      */
@@ -95,12 +126,45 @@ public final class CLContext {
 
         return null;
     }
+*/
 
+    /**
+     * Returns all devices associated with this CLContext.
+     */
     public CLDevice[] getCLDevices() {
 
+        if(devices == null) {
+
+            int sizeofDeviceID = 8; // TODO doublechek deviceID size on 32 bit systems
+
+            long[] longBuffer = new long[1];
+
+            int ret;
+            ret = cl.clGetContextInfo(contextID, CL.CL_CONTEXT_DEVICES, 0, null, longBuffer, 0);
+            checkForError(ret, "can not enumerate devices");
+
+            ByteBuffer deviceIDs = ByteBuffer.allocate((int)longBuffer[0]).order(ByteOrder.nativeOrder());
+
+            ret = cl.clGetContextInfo(contextID, CL.CL_CONTEXT_DEVICES, deviceIDs.capacity(), deviceIDs, null, 0);
+            checkForError(ret, "can not enumerate devices");
+
+            devices = new CLDevice[deviceIDs.capacity()/sizeofDeviceID];
+            for (int i = 0; i < devices.length; i++)
+                devices[i] = new CLDevice(cl, deviceIDs.getLong()); // TODO doublechek deviceID size on 32 bit systems
+
+        }
+
+        return devices;
     }
-*/
 
+    CLDevice getCLDevices(long dID) {
+        CLDevice[] deviceArray = getCLDevices();
+        for (int i = 0; i < deviceArray.length; i++) {
+            if(dID == deviceArray[i].deviceID)
+                return deviceArray[i];
+        }
+        return null;
+    }
 
     /**
      * Lists all available OpenCL implementaitons.
@@ -111,14 +175,12 @@ public final class CLContext {
         int[] intBuffer = new int[1];
         // find all available OpenCL platforms
         int ret = cl.clGetPlatformIDs(0, null, 0, intBuffer, 0);
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "can not enumerate platforms");
+        checkForError(ret, "can not enumerate platforms");
 
         // receive platform ids
         long[] platformId = new long[intBuffer[0]];
         ret = cl.clGetPlatformIDs(platformId.length, platformId, 0, null, 0);
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "can not enumerate platforms");
+        checkForError(ret, "can not enumerate platforms");
 
         CLPlatform[] platforms = new CLPlatform[platformId.length];
 
@@ -135,4 +197,5 @@ public final class CLContext {
         return cl;
     }
 
+
 }
diff --git a/src/com/mbien/opencl/CLDevice.java b/src/com/mbien/opencl/CLDevice.java
index ba20f4b2..d9f643ce 100644
--- a/src/com/mbien/opencl/CLDevice.java
+++ b/src/com/mbien/opencl/CLDevice.java
@@ -7,6 +7,8 @@ import java.util.HashSet;
 import java.util.Scanner;
 import java.util.Set;
 
+import static com.mbien.opencl.CLException.*;
+
 /**
  * 
  * @author Michael Bien
@@ -34,7 +36,7 @@ public final class CLDevice {
          */
         DEFAULT(CL.CL_DEVICE_TYPE_DEFAULT);
 
-         /**
+        /**
          * Value of wrapped OpenCL device type.
          */
         public final int CL_TYPE;
@@ -160,8 +162,7 @@ public final class CLDevice {
 
         int ret = cl.clGetDeviceInfo(deviceID, key, bb.capacity(), bb, null, 0);
 
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "can not receive device info");
+        checkForError(ret, "can not receive device info");
 
         return bb.getLong();
     }
@@ -173,8 +174,7 @@ public final class CLDevice {
 
         int ret = cl.clGetDeviceInfo(deviceID, key, bb.capacity(), bb, longBuffer, 0);
         
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "can not receive device info string");
+        checkForError(ret, "can not receive device info string");
 
         return new String(bb.array(), 0, (int)longBuffer[0]);
         
@@ -183,9 +183,10 @@ public final class CLDevice {
 
     @Override
     public String toString() {
-        return "CLPlatform [name:" + getName()
-                        + " type:" + getType()
-                        + " profile: " + getProfile()+"]";
+        return "CLDevice [id: " + deviceID
+                      + " name: " + getName()
+                      + " type: " + getType()
+                      + " profile: " + getProfile()+"]";
     }
 
     @Override
diff --git a/src/com/mbien/opencl/CLException.java b/src/com/mbien/opencl/CLException.java
index f439e89b..5806d41d 100644
--- a/src/com/mbien/opencl/CLException.java
+++ b/src/com/mbien/opencl/CLException.java
@@ -22,6 +22,12 @@ public class CLException extends RuntimeException {
         super(identifyError(error) + ": " + message);
     }
 
+    public static final void checkForError(int status, String message) {
+        if(status != CL.CL_SUCCESS)
+            throw new CLException(status, message);
+    }
+
+
     private static final String identifyError(int error) {
 
         switch (error) {
diff --git a/src/com/mbien/opencl/CLPlatform.java b/src/com/mbien/opencl/CLPlatform.java
index 252872dc..a717564a 100644
--- a/src/com/mbien/opencl/CLPlatform.java
+++ b/src/com/mbien/opencl/CLPlatform.java
@@ -1,7 +1,7 @@
 package com.mbien.opencl;
 
 import java.nio.ByteBuffer;
-
+import static com.mbien.opencl.CLException.*;
 /**
  *
  * @author Michael Bien
@@ -29,13 +29,11 @@ public final class CLPlatform {
 
         //find all devices
         int ret = cl.clGetDeviceIDs(platformID, CL.CL_DEVICE_TYPE_ALL, 0, null, 0, intBuffer, 0);
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "error while enumerating devices");
+        checkForError(ret, "error while enumerating devices");
 
         long[] deviceIDs = new long[intBuffer[0]];
         ret = cl.clGetDeviceIDs(platformID, CL.CL_DEVICE_TYPE_ALL, deviceIDs.length, deviceIDs, 0, null, 0);
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "error while enumerating devices");
+        checkForError(ret, "error while enumerating devices");
 
         CLDevice[] devices = new CLDevice[deviceIDs.length];
 
@@ -83,8 +81,7 @@ public final class CLPlatform {
         ByteBuffer bb = ByteBuffer.allocate(512);
 
         int ret = cl.clGetPlatformInfo(platformID, key, bb.capacity(), bb, longBuffer, 0);
-        if(CL.CL_SUCCESS != ret)
-            throw new CLException(ret, "can not receive info string");
+        checkForError(ret, "can not receive info string");
 
         return new String(bb.array(), 0, (int)longBuffer[0]);
     }
diff --git a/src/com/mbien/opencl/CLProgram.java b/src/com/mbien/opencl/CLProgram.java
new file mode 100644
index 00000000..e93e7246
--- /dev/null
+++ b/src/com/mbien/opencl/CLProgram.java
@@ -0,0 +1,197 @@
+package com.mbien.opencl;
+
+import java.nio.ByteBuffer;
+import java.nio.ByteOrder;
+import static com.mbien.opencl.CLException.*;
+
+/**
+ *
+ * @author Michael Bien
+ */
+public class CLProgram {
+    
+    private final CLContext context;
+    private final long programID;
+    private final CL cl;
+    
+    public enum Status {
+        
+        BUILD_SUCCESS(CL.CL_BUILD_SUCCESS), 
+        BUILD_NONE(CL.CL_BUILD_NONE), 
+        BUILD_IN_PROGRESS(CL.CL_BUILD_IN_PROGRESS),
+        BUILD_ERROR(CL.CL_BUILD_ERROR);
+                
+        /**
+         * Value of wrapped OpenCL device type.
+         */
+        public final int CL_BUILD_STATUS;
+
+        private Status(int CL_BUILD_STATUS) {
+            this.CL_BUILD_STATUS = CL_BUILD_STATUS;
+        }
+        
+        public static Status valueOf(int clBuildStatus) {
+            switch(clBuildStatus) {
+                case(CL.CL_BUILD_SUCCESS):
+                    return BUILD_SUCCESS;
+                case(CL.CL_BUILD_NONE):
+                    return BUILD_NONE;
+                case(CL.CL_BUILD_IN_PROGRESS):
+                    return BUILD_IN_PROGRESS;
+                case(CL.CL_BUILD_ERROR):
+                    return BUILD_ERROR;
+// is this a standard state?
+//              case (CL.CL_BUILD_PROGRAM_FAILURE):
+//                    return BUILD_PROGRAM_FAILURE;
+            }
+            return null;
+        }
+    }
+
+    CLProgram(CLContext context, String src, long contextID) {
+        
+        this.cl = context.cl;
+        this.context = context;
+
+        int[] intArray = new int[1];
+        // Create the program
+        programID = cl.clCreateProgramWithSource(contextID, 1, new String[] {src}, new long[]{src.length()}, 0, intArray, 0);
+        checkForError(intArray[0], "can not create program with source");
+    }
+
+
+    /**
+     * Builds this program for all devices accosiated with the context and implementation specific build options.
+     * @return this
+     */
+    public CLProgram build() {
+        build(null, null);
+        return this;
+    }
+
+    /**
+     * Builds this program for the given devices and with the specified build options.
+     * @return this
+     * @param devices A list of devices this program should be build on or null for all devices of its context.
+     */
+    public CLProgram build(CLDevice[] devices, String options) {
+
+        long[] deviceIDs = null;
+        if(devices != null) {
+            deviceIDs = new long[devices.length];
+            for (int i = 0; i < deviceIDs.length; i++) {
+                deviceIDs[i] = devices[i].deviceID;
+            }
+        }
+
+        // Build the program
+        int ret = cl.clBuildProgram(programID, deviceIDs, options, null, null);
+        checkForError(ret, "error building program");
+
+        return this;
+    }
+
+    /**
+     * Releases this program.
+     * @return this
+     */
+    public CLProgram release() {
+        int ret = cl.clReleaseProgram(programID);
+        checkForError(ret, "can not release program");
+        context.programReleased(this);
+
+        return this;
+    }
+
+    /**
+     * Returns all devices associated with this program.
+     */
+    public CLDevice[] getCLDevices() {
+
+        long[] longArray = new long[1];
+        int ret = cl.clGetProgramInfo(programID, CL.CL_PROGRAM_DEVICES, 0, null, longArray, 0);
+        checkForError(ret, "on clGetProgramInfo");
+
+        ByteBuffer bb = ByteBuffer.allocate((int) longArray[0]).order(ByteOrder.nativeOrder());
+        ret = cl.clGetProgramInfo(programID, CL.CL_PROGRAM_DEVICES, bb.capacity(), bb, null, 0);
+        checkForError(ret, "on clGetProgramInfo");
+
+        int count = bb.capacity() / 8; // TODO sizeof cl_device
+        CLDevice[] devices = new CLDevice[count];
+        for (int i = 0; i < count; i++) {
+            devices[i] = context.getCLDevices(bb.getLong());
+        }
+
+        return devices;
+
+    }
+
+    public String getBuildLog(CLDevice device) {
+        return getBuildInfoString(device.deviceID, CL.CL_PROGRAM_BUILD_LOG);
+    }
+
+    public Status getBuildStatus(CLDevice device) {
+        int clStatus = getBuildInfoInt(device.deviceID, CL.CL_PROGRAM_BUILD_STATUS);
+        return Status.valueOf(clStatus);
+    }
+
+    /**
+     * Returns the source code of this program. Note: sources are not cached, each call of this method calls into OpenCL.
+     */
+    public String getSource() {
+        return getProgramInfoString(CL.CL_PROGRAM_SOURCE);
+    }
+
+    // TODO binaries, serialization, program build options
+
+    private final String getBuildInfoString(long device, int flag) {
+
+        long[] longArray = new long[1];
+
+        int ret = cl.clGetProgramBuildInfo(programID, device, flag, 0, null, longArray, 0);
+        checkForError(ret, "on clGetProgramBuildInfo");
+
+        ByteBuffer bb = ByteBuffer.allocate((int)longArray[0]).order(ByteOrder.nativeOrder());
+
+        ret = cl.clGetProgramBuildInfo(programID, device, flag, bb.capacity(), bb, null, 0);
+        checkForError(ret, "on clGetProgramBuildInfo");
+
+        return new String(bb.array(), 0, (int)longArray[0]);
+    }
+
+    private final String getProgramInfoString(int flag) {
+
+        long[] longArray = new long[1];
+
+        int ret = cl.clGetProgramInfo(programID, flag, 0, null, longArray, 0);
+        checkForError(ret, "on clGetProgramInfo");
+
+        ByteBuffer bb = ByteBuffer.allocate((int)longArray[0]).order(ByteOrder.nativeOrder());
+
+        ret = cl.clGetProgramInfo(programID, flag, bb.capacity(), bb, null, 0);
+        checkForError(ret, "on clGetProgramInfo");
+
+        return new String(bb.array(), 0, (int)longArray[0]);
+    }
+
+//    private int getProgramInfoInt(int flag) {
+//
+//        ByteBuffer bb = ByteBuffer.allocate(4).order(ByteOrder.nativeOrder());
+//
+//        int ret = cl.clGetProgramInfo(programID, flag, bb.capacity(), bb, null, 0);
+//        checkForError(ret, "");
+//
+//        return bb.getInt();
+//    }
+    
+    private int getBuildInfoInt(long device, int flag) {
+
+        ByteBuffer bb = ByteBuffer.allocate(4).order(ByteOrder.nativeOrder());
+
+        int ret = cl.clGetProgramBuildInfo(programID, device, flag, bb.capacity(), bb, null, 0);
+        checkForError(ret, "error on clGetProgramBuildInfo");
+
+        return bb.getInt();
+    }
+
+}
diff --git a/test/com/mbien/opencl/JOCLTest.java b/test/com/mbien/opencl/JOCLTest.java
index 15356ae2..05b1d28d 100644
--- a/test/com/mbien/opencl/JOCLTest.java
+++ b/test/com/mbien/opencl/JOCLTest.java
@@ -16,6 +16,20 @@ import static java.lang.System.*;
  */
 public class JOCLTest {
 
+    private final static String programSource =
+              " // OpenCL Kernel Function for element by element vector addition                                                \n"
+            + "__kernel void VectorAdd(__global const int* a, __global const int* b, __global int* c, int iNumElements) {       \n"
+            + "    // get index into global data array                                                                          \n"
+            + "    int iGID = get_global_id(0);                                                                                 \n"
+            + "    // bound check (equivalent to the limit on a 'for' loop for standard/serial C code                           \n"
+            + "    if (iGID >= iNumElements)  {                                                                                 \n"
+            + "        return;                                                                                                  \n"
+            + "    }                                                                                                            \n"
+            + "    // add the vector elements                                                                                   \n"
+            + "    c[iGID] = a[iGID] + b[iGID];                                                                                 \n"
+            + "    //c[iGID] = iGID;                                                                                            \n"
+            + "}                                                                                                                \n";
+
     public JOCLTest() {
     }
 
@@ -119,8 +133,7 @@ public class JOCLTest {
         long context = cl.clCreateContextFromType(null, 0, CL.CL_DEVICE_TYPE_ALL, null, null, null, 0);
         out.println("context handle: "+context);
 
-        // TODO fix gluegen bug: array-buffer mixing... bb is a noop
-        ret = cl.clGetContextInfo(context, CL.CL_CONTEXT_DEVICES, 0, bb, longArray, 0);
+        ret = cl.clGetContextInfo(context, CL.CL_CONTEXT_DEVICES, 0, null, longArray, 0);
         checkError("on clGetContextInfo", ret);
 
         int sizeofLong = 8; // TODO sizeof long...
@@ -154,23 +167,9 @@ public class JOCLTest {
         long devDst  = cl.clCreateBuffer(context, CL.CL_MEM_WRITE_ONLY, BufferFactory.SIZEOF_INT * globalWorkSize, null, intArray, 0);
         checkError("on clCreateBuffer", intArray[0]);
 
-        String src =
-              " // OpenCL Kernel Function for element by element vector addition                                                \n"
-            + "__kernel void VectorAdd(__global const int* a, __global const int* b, __global int* c, int iNumElements) {       \n"
-            + "    // get index into global data array                                                                          \n"
-            + "    int iGID = get_global_id(0);                                                                                 \n"
-            + "    // bound check (equivalent to the limit on a 'for' loop for standard/serial C code                           \n"
-            + "    if (iGID >= iNumElements)  {                                                                                 \n"
-            + "        return;                                                                                                  \n"
-            + "    }                                                                                                            \n"
-            + "    // add the vector elements                                                                                   \n"
-            + "    c[iGID] = a[iGID] + b[iGID];                                                                                 \n"
-            + "    //c[iGID] = iGID;                                                                                            \n"
-            + "}                                                                                                                \n";
-
 
         // Create the program
-        long program = cl.clCreateProgramWithSource(context, 1, new String[] {src}, new long[]{src.length()}, 0, intArray, 0);
+        long program = cl.clCreateProgramWithSource(context, 1, new String[] {programSource}, new long[]{programSource.length()}, 0, intArray, 0);
         checkError("on clCreateProgramWithSource", intArray[0]);
 
         // Build the program
@@ -186,7 +185,7 @@ public class JOCLTest {
         ret = cl.clGetProgramInfo(program, CL.CL_PROGRAM_SOURCE, 0, bb, longArray, 0);
         checkError("on clGetProgramInfo CL_PROGRAM_SOURCE", ret);
         out.println("program source length (cl): "+longArray[0]);
-        out.println("program source length (java): "+src.length());
+        out.println("program source length (java): "+programSource.length());
 
         bb.rewind();
         ret = cl.clGetProgramInfo(program, CL.CL_PROGRAM_SOURCE, bb.capacity(), bb, null, 0);
@@ -203,8 +202,7 @@ public class JOCLTest {
         assertEquals("build status", CL.CL_BUILD_SUCCESS, bb.getInt(0));
 
         // Read build log
-        // TODO fix gluegen bug: array-buffer mixing... bb is a noop
-        ret = cl.clGetProgramBuildInfo(program, firstDeviceID, CL.CL_PROGRAM_BUILD_LOG, 0, bb, longArray, 0);
+        ret = cl.clGetProgramBuildInfo(program, firstDeviceID, CL.CL_PROGRAM_BUILD_LOG, 0, null, longArray, 0);
         checkError("on clGetProgramBuildInfo2", ret);
         out.println("program log length: " + longArray[0]);
 
@@ -328,9 +326,9 @@ public class JOCLTest {
     }
 
     @Test
-    public void highLevelTest() {
+    public void highLevelTest1() {
         
-        out.println(" - - - highLevelTest - - - ");
+        out.println(" - - - highLevelTest; contextless - - - ");
 
         CLPlatform[] clPlatforms = CLContext.listCLPlatforms();
 
@@ -359,14 +357,46 @@ public class JOCLTest {
         }
 
 
-        CLContext ctx = CLContext.create();
-//        CLDevice device = ctx.getMaxFlopsDevice();
-//        out.println("max FLOPS device: " + device);
-        ctx.release();
     }
 
 
-    
+    @Test
+    public void highLevelTest2() {
+
+        out.println(" - - - highLevelTest - - - ");
+
+        CLContext context = CLContext.create();
+
+        CLDevice[] contextDevices = context.getCLDevices();
+
+        out.println("context devices:");
+        for (CLDevice device : contextDevices) {
+            out.println("   "+device.toString());
+        }
+
+        CLProgram program = context.createProgram(programSource).build();
+
+        CLDevice[] programDevices = program.getCLDevices();
+
+        assertEquals(contextDevices.length, programDevices.length);
+
+        out.println("program devices:");
+        for (CLDevice device : programDevices) {
+            out.println("   "+device.toString());
+            out.println("   build log: "+program.getBuildLog(device));
+            out.println("   build status: "+program.getBuildStatus(device));
+        }
+
+        out.println("source:\n"+program.getSource());
+
+        assertTrue(1 == context.getPrograms().size());
+        program.release();
+        assertTrue(0 == context.getPrograms().size());
+
+//        CLDevice device = ctx.getMaxFlopsDevice();
+//        out.println("max FLOPS device: " + device);
+        context.release();
+    }
 
 
     private final int roundUp(int groupSize, int globalSize) {