ifca-advanced-computing/opencl

Haskell high-level wrapper for OpenCL

OpenCL

A high-level OpenCL library for Haskell.

Based on the OpenCLRaw package by J.R. Heard.

By Luis Cabellos at IFCA

Installation

Requirements: c2hs must be installed. (Try cabal install c2hs.)

With the usual commands to install as a user library:

cabal install --user

Programs using the library must link against OpenCL; for example, by passing -lOpenCL to GHC.

About versioning

OpenCL module uses Package Version Policy:

http://www.haskell.org/haskellwiki/Package_versioning_policy

But It differs in the A version number. It use OpenCL API version as A number, so 1.0.3.0 correspond to A=1.0=OpenCL API version 1.0, B=3 and C=0. The major version number is 1.0.3

Optional Requisites

Some OpenCL libraries require additional NUMA libraries. For instance, on Ubuntu 11.04:

sudo apt-get install libnuma1 libnuma-dev

Example

There is an simple working example in the examples folder. You can create an executable using:

ghc --make examples/example01.hs

Using ghci

It’s possible to use GHCi with OpenCL, e.g.:

ghci examples/example01.hs

opencl's Issues

Compilation failure on OS X El Capitan

I failed to compile OpenCL-1.0.3.4 on Mac OS X El Capitan (10.11.4) with the following error.

Preprocessing library OpenCL-1.0.3.4...
c2hs: C header contains errors:

/System/Library/Frameworks/OpenGL.framework/Headers/CGLDevice.h:19: (column 28) [ERROR]  >>> Syntax error !
  The symbol `CGLGetShareGroup' does not fit here.

It seems that the __nullable annotation confused c2hs.
Here are some code snippet that might be relevant to the problem.

CGLDevice.h:

15  OPENGL_ASSUME_NONNULL_BEGIN
16  
17  typedef struct CGLShareGroupRec *CGLShareGroupObj OPENGL_AVAILABLE(10_6);
18  
19  CGLShareGroupObj OPENGL_NULLABLE CGLGetShareGroup(CGLContextObj ctx) OPENGL_AVAILABLE(10_6);
20  

CGLTypes.h:

14  #if __has_feature(assume_nonnull)
15  #define OPENGL_ASSUME_NONNULL_BEGIN _Pragma("clang assume_nonnull begin")
16  #define OPENGL_ASSUME_NONNULL_END _Pragma("clang assume_nonnull end")
17  #else
18  #define OPENGL_ASSUME_NONNULL_BEGIN
19  #define OPENGL_ASSUME_NONNULL_END
20  #endif
21  
22  #if __has_feature(nullability)
23  #define OPENGL_NULLABLE __nullable
24  #define OPENGL_NONNULL __nonnull
25  #else
26  #define OPENGL_NULLABLE
27  #define OPENGL_NONNULL
28  #endif

clGetDeviceIDs throws error instead of returning []

The docs for clGetDeviceIDs say

-- | Obtain the list of devices available on a platform. Returns the list if
-- the function is executed successfully. Otherwise it returns the empty list
-- if platform is not a valid platform or no OpenCL devices that matched
-- device_type were found.

The implementation calls whenSuccess, and throws CLError if no matching devices are found.

Should the implementation or the documentation be changed?

Define clCreateProgramWithBinary function

Define clCreateProgramWithBinary function.

Change functions return when integral

There is a lot of functions having signature:

clGetXXXYYY :: CLXXXID -> IO CLuint

Should the signature be with Integral a instead? Like:

clGetXXXYYY :: Integral a => CLXXXID -> IO a

Think about it,

Fix problem with newlines on clCreateProgramWithSource

OpenCL preprocessor actually wants newlines to be there. So current implementation fails because we remove it.

See:
ethorsoe/OpenCLWrappers@242a21e

Don't remove newlines from source parameter.

Think about get[Type]Info naming

In OpenCL package the clGet[Type]Info functions are splitted in several clGet[Type][specificInfo] functions. is it better naming convention use clGet[Type]Info_[specificInfo]?

e.g, c call:

clGetProgramInfo( aProgram, CL_PROGRAM_CONTEXT, sizeof(cxt), (void*)&ctx, &size )

Should be in Haskell:

1. ctx <- clGetProgramContext aProgram
2. ctx <- clGetProgramInfo_ProgramContext aProgram

Option 1 is the current one, option 2 has the advantage of better searching in documentation (it contains the original name).

Change to BSD3 License

The GPL is known to cause problems with Haskell code due to cross module inlining. [need ref]

OpenCL package will use BSD3 License in next iteration.

Documentation for clEnqueueReadImage is confusing!

I think the documentation for clEnqueueReadImage and clEnqueueWriteImage are a bit confusing. For example, for read image it says:

Enqueues a command to read from a 2D or 3D image object to host memory. 

But for write image it says:

Enqueues a command to write from a 2D or 3D image object to host memory. 

This confusion appears in several places, such as in the Haddock documentation for what the Ptr () actually refers to. In clEnqueueReadImage, it says that pointer is

The pointer to a buffer in host memory where image data is to be read from. 

Which makes no sense, I think, since I want to read from device memory into host memory?

Anyway I'm confused.

Thanks!

Giving compile options to clBuildProgram

Sometimes you have to generate macros automatically by the host program and inject them in the OpenCL-C code before it is compiled. In the C++-API therefore you can give the clBuildProgram-call a pointer to an option string. In the current version of the Haskell-API you can simulate this behaviour by prepending this string to the one you give the Haskell-call to compile. Unfortunately if you do this, the line number increases and if you receive an error message by the cl-compiler the line numbers it tells are too high.

Therefore it would be nice to be able to give the Haskell-call an additional (possibly empty) String and change it's signature to:
clBuildProgram :: CLProgram -> [CLDeviceID] -> String -> String -> IO ()

clSetKernelArg cannot specify size of local memory arrays

I asked on stack overflow how to use clSetKernelArg with locals but someone suggested I report it as a bug instead. It seems impossible to use the exported clSetKernelArg function to specify local arrays. I really like that it determines the size and pointer for the given Storable automatically, but unfortunately this doesn't work for the intended usage with locals, where you are supposed to specify the size along with a null-pointer. Passing nullPtr does not work, since there is no way to specify the size.

Possibly suggestion: don't wrap clSetKernelArg with sizeOf and with, instead make the user do it. Perhaps provide a higher-level function doing this automatically, with a different name like clSetKernelArgStorable.

Alternative suggestion: additionally provide clSetKernelArgLocal specifically for using with locals.

I like the second suggestion better, but the first suggestion is more in keeping with the original OpenCL API, so I'm not sure which idea is best.

Can opencl install on Microsoft windows 8

I try on cmd with "cabal install opencl".
But it failure. Input from command line show below.

https://docs.google.com/file/d/0B02yMjOZsqqFLWtRMzVtZU90dDQ/edit?usp=sharing

Sorry for my bad english.

Upload 1.0.3.4 to hackage

Can you please upload the latest version to hackage? I am trying to release some code that requires it.

Cabal install depends on c2hs

When I run cabal install opencl on Mac OSX 10.7.5, I get the following error:

Resolving dependencies...
Downloading OpenCL-1.0.3.4...
Configuring OpenCL-1.0.3.4...
cabal: The program c2hs is required but it could not be found.
Failed to install OpenCL-1.0.3.4
cabal: Error: some packages failed to install:
OpenCL-1.0.3.4 failed during the configure step. The exception was:
ExitFailure 1

Running cabal install c2hs followed by cabal install opencl worked perfectly.

It looks like the .cabal file lists c2hs in build-tools but not build-depends, and for whatever reason, cabal install doesn't verify that c2hs is installed. There seems to be some documentation on these options here. I tried to figure out the right thing to do but am not sure.

Use stdcall instead of ccall (at least on Windows)

For reference: http://stackoverflow.com/questions/7391634/cant-link-opencl-on-windows-with-ghc

unknown symbol `clGetDeviceIDs' when trying to use OpenCL

When I'm trying to build any code using OpenCL with:

ghc -lOpenCL --make Test.hs

I'm getting:

ghc: /home/tomek/.cabal/lib/OpenCL-1.0.3.0/ghc-7.4.1/HSOpenCL-1.0.3.0.o: unknown symbol `clGetDeviceIDs'

It's the same for cabal-ghci in project using OpenCL, still, it works with runhaskell -lOpenCL Test.hs

it looks like adding "extra-libraries: OpenCL" in Library section of cabal file fixes this issue.

Add testing

Add test in the cabal package

Define Image Buffer functions

Define functions:

• clCreateImage2D
• clCreateImage3D
• clGetSupportedImageFormats
• clEnqueueReadImage
• clEnqueueWriteImage
• clEnqueueCopyImage
• clEnqueueCopyImageToBuffer
• clEnqueueCopyBufferToImage
• clEnqueueMapBuffer
• clEnqueueMapImage
• clEnqueueUnmapMemObject
• clGetImageInfo

Testing V1.0.0.2

Tested on Ubuntu 11.04 64 bits.

While following the steps described in the README file, two dependency problems arised:

• running "runhaskell Setup configure --user

  277 spike@bebop:~/workspace/haskell/opencl% runhaskell Setup configure --user
  Configuring OpenCL-1.0.0.2...
  Setup: The program c2hs is required but it could not be found.

SOLUTION: install c2hs

• at compiling example.hs (warning)

  280 spike@bebop:~/workspace/haskell/opencl% ghc --make -lOpenCL examples/example01.hs
  [1 of 1] Compiling Main ( examples/example01.hs, examples/example01.o )
  Linking examples/example01 ...
  /usr/bin/ld: warning: libnuma.so.1, needed by /usr/lib64/libOpenCL.so, not found (try using -rpath or -rpath-link)

SOLUTION: install libnuma1 libnuma-dev

After solving this two small problems, everything works OK.

Better install proccess with configure

Add a better install proccess with configure that check for opencl.h include or use a default one.

OpenCL fails to build on OSX Mavericks 10.9.1: c2hs cannot parse resource.h

On my machine, I cannot seem to be able to install OpenCL, as the following error appears:

Georges-MacBook-Pro:Downloads georgeburton$ cabal install opencl
Resolving dependencies...
Configuring OpenCL-1.0.3.4...
Building OpenCL-1.0.3.4...
Preprocessing library OpenCL-1.0.3.4...
c2hs: C header contains errors:

/usr/include/sys/resource.h:349: (column 75) [ERROR]  >>> Syntax error !
The symbol `=' does not fit here.

Failed to install OpenCL-1.0.3.4
cabal: Error: some packages failed to install:
OpenCL-1.0.3.4 failed during the building phase. The exception was:
ExitFailure 1

My GHC version is:
The Glorious Glasgow Haskell Compilation System, version 7.6.3

And my cabal version is:
cabal-install version 1.18.0.2
using version 1.18.1.2 of the Cabal library

Calling `getPlatformIDs` throw `CL_PLATFORM_NOT_FOUND_KHR`

Hi,

I'm running GHC 7.8.3 an 64bit archlinux, I have also installed opencl-nvidia-340.32 and cuda-6.5.14.

I tried to link against:
libOpenCL.so.1 => /opt/cuda/lib64/libOpenCL.so.1 (0x00007f36ed4c0000).
libOpenCL.so.1 => /usr/lib/libOpenCL.so.1 (0x00007f6370fee000) (supposedly opencl-nivida)

When trying to run the examples I'm getting the following exception:
CL_PLATFORM_NOT_FOUND_KHR

My graphic card is a GeForce 9800 GT.

It's because of some incompatibility I suppose?

Thanks for your help

Ptr content not complete after call to clGetProgramBuildLog

Hi,

I want to read the compile output for my OpenCL application (using intel ICD). Now, the following source code does not work.

import Control.Parallel.OpenCL
import Foreign( castPtr, nullPtr, sizeOf )
import Foreign.C.Types( CFloat )
import Foreign.Marshal.Array( newArray, peekArray )
import Control.Exception

programSource :: String
programSource = "I won't compile!"

main :: IO ()
main = do
  -- Initialize OpenCL
  (platform:_) <- clGetPlatformIDs
  (dev:_) <- clGetDeviceIDs platform CL_DEVICE_TYPE_ALL
  context <- clCreateContext [CL_CONTEXT_PLATFORM platform] [dev] print
  q <- clCreateCommandQueue context dev []
  
  -- Initialize Kernel
  program <- clCreateProgramWithSource context programSource
  clBuildProgram program [dev] "" `catch` (\e -> case (e ::CLError) of 
              CL_BUILD_PROGRAM_FAILURE -> do 
                putStrLn "Building OpenCL program failed. Build Status:"
                --errMsg <- clGetProgramBuildLog program dev
                --putStrLn errMsg
                putStrLn <$> clGetProgramBuildLog program dev
                return ())
  putStrLn "done"

If I manually enable the lines

--errMsg <- clGetProgramBuildLog program dev
--putStrLn errMsg

and disable the following it works.

Maybe this is more related to FFI but I'm not sure.

I'm using ghc-8.10.3 and Haskell-OpenCL 1.0.3.4

Any ideas?

clGetKernelArgInfo functions

These are OpenCL 1.2. This would entail adding:

clGetKernelArgTypeName :: CLKernel -> Int -> IO String
clGetKernelArgName :: CLKernel -> Int -> IO String
clGetKernelArgAddressQualifier :: CLKernel -> Int -> IO CLKernelArgAddressQualifier
clGetKernelArgAccessQualifier :: CLKernel -> Int -> IO CLKernelArgAccessQualifier
clGetKernelArgTypeQualifier :: CLKernel -> Int -> IO CLKernelArgTypeQualifier
And the enumerations for the latter 3.
data CLKernelArgAddressQualifier =
CL_KERNEL_ARG_ADDRESS_GLOBAL
| ...
... etc

I wouldn't mind implementing this and setting up a pull request (this would be my first).

Integration with OpenGL

OpenCL have integration with OpenGL. It would be nice to export it (basing on Haskell opengl bindings) as well.

Module name suggestion

System.GPU.OpenCL seems a bit overly specific. I know OpenCL was intended for GPGPU programming but Intel also provides drivers for targeting multicore CPUs.

What about these alternatives:

• System.Compute.OpenCL
• System.Parallel.OpenCL
• System.OpenCL
• Control.Parallel.OpenCL

I got the last one from here: http://stackoverflow.com/questions/6647481/whats-the-best-hierarchical-module-path-for-an-opencl-haskell-library

I think you asked that question on SO? My favorite is Control.Parallel.OpenCL.

evaluate the use of #fun c2hs

Evaluate the use of #fun macros in *.chs files for a better binding.

Define clGetContextProperties function

Define clGetContextProperties function.

OpenCL MAN

Error-handling is suboptimal

I'm interesting in using this library, but the IMHO suboptimal and non-idiomatic error handling style is putting me off a bit.

It would be much nicer if, instead of IO (Either CLError a), every OpenCL action was IO a, and an exception type was used, by simply deriving an instance of Typeable for CLError and adding an empty instance Exception CLError.

This allows handling errors just as simply as with the current API; this code:

do result <- clAction
   case result of
     Left err -> handle error
     Right a -> ...

becomes

do a <- clAction `catch` \(CLException err) ->
     handle error
   ...

and the simpler case, in which errors are not handled at this point (bubbling up to a higher handler or just making the program fail with an error message if there is no reasonable recovery path) becomes a lot simpler; the code:

either (fail . show) id <$> clAction

(à la your myTry -- but it's bad practice to use error e as a monadic value; fail e or return (error e) are better, but the latter only triggers when you force the value returned)

becomes simply:

clAction

And if the original error-returning behaviour is desired, it can be recreated as simply as this:

try clAction

(with Control.Exception.try)

Thanks for taking the time to consider this request! :)

Cabal file lists incorrect Issue Tracker URL

See the Hackage page for this project.

Define clEnqueueNativeKernel function

Define clEnqueueNativeKernel function.

Segmentation fault when using CUChar/uint8

While playing with this library (first time I use OpenCL), I got into an issue resulting in crashers (segmentation faults) when using CUChar/uint8. Here's what happens:

I try to run the first example (example01.hs), but altered it slightly to use 8-bit integers instead of floats, here's the diff:

--- example01.hs    2011-12-29 00:13:44.127347130 +0100
+++ example01bis.hs 2011-12-29 00:19:21.372373866 +0100
@@ -31,11 +31,11 @@
 -}
 import System.GPU.OpenCL
 import Foreign( castPtr, nullPtr, sizeOf )
-import Foreign.C.Types( CFloat )
+import Foreign.C.Types( CUChar )
 import Foreign.Marshal.Array( newArray, peekArray )

 programSource :: String
-programSource = "__kernel void duparray(__global float *in, __global float *out ){\n  int id = get_global_id(0);\n  out[id] = 2*in[id];\n}"
+programSource = "__kernel void duparray(__global uint8 *in, __global uint8 *out ){\n  int id = get_global_id(0);\n  out[id] = 2*in[id];\n}"

 main :: IO ()
 main = do
@@ -51,8 +51,8 @@
   kernel <- clCreateKernel program "duparray"

   -- Initialize parameters
-  let original = [0 .. 20] :: [CFloat]
-      elemSize = sizeOf (0 :: CFloat)
+  let original = [0 .. 20] :: [CUChar]
+      elemSize = sizeOf (0 :: CUChar)
       vecSize = elemSize * length original
   putStrLn $ "Original array = " ++ show original
   input  <- newArray original

Running the float example works as expected:

[nicolas@tau mvmul]$ ghc --make -lOpenCL -fforce-recomp example01.hs
[1 of 1] Compiling Main             ( example01.hs, example01.o )
Linking example01 ...
[nicolas@tau mvmul]$ ./example01
Original array = [0.0,1.0,2.0,3.0,4.0,5.0,6.0,7.0,8.0,9.0,10.0,11.0,12.0,13.0,14.0,15.0,16.0,17.0,18.0,19.0,20.0]
Result array = [0.0,2.0,4.0,6.0,8.0,10.0,12.0,14.0,16.0,18.0,20.0,22.0,24.0,26.0,28.0,30.0,32.0,34.0,36.0,38.0,40.0]

and doesn't fail, even upon repeated execution.

The altered version crashes:

[nicolas@tau mvmul]$ ghc --make -lOpenCL -fforce-recomp example01bis.hs
[1 of 1] Compiling Main             ( example01bis.hs, example01bis.o )
Linking example01bis ...
[nicolas@tau mvmul]$ ./example01bis 
Original array = [0,1,2,3,4,5,6,7,8,9,10,11,12,13,14,15,16,17,18,19,20]
Result array = [0,2,4,6,8,10,12,14,16,18,20,22,24,26,28,30,32,34,36,38,40]
*** glibc detected *** ./example01bis: corrupted double-linked list: 0x0000000001ef6d10 ***
======= Backtrace: =========
/lib64/libc.so.6[0x314be7c2d6]
/lib64/libc.so.6[0x314be7c7a9]
/lib64/libc.so.6[0x314be7d3c9]
/home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so(+0xdfa26d)[0x7fcde48e426d]
/home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so(+0xdfb51e)[0x7fcde48e551e]
/home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so(+0xe5d97d)[0x7fcde494797d]
/home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so(+0xe5d9a5)[0x7fcde49479a5]
/home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so(+0x1830b9)[0x7fcde3c6d0b9]
/lib64/libc.so.6[0x314be39931]
/lib64/libc.so.6[0x314be399b5]
./example01bis[0x4d21a8]
./example01bis[0x4d21cd]
./example01bis[0x4daab1]
./example01bis[0x4daafe]
/lib64/libc.so.6(__libc_start_main+0xed)[0x314be2169d]
./example01bis[0x406b45]
======= Memory map: ========
(snip)

Loading the coredump using gdb to retrieve symbol information gives:

#0  0x000000314be36285 in __GI_raise (sig=6)
    at ../nptl/sysdeps/unix/sysv/linux/raise.c:64
#1  0x000000314be37b9b in __GI_abort () at abort.c:91
#2  0x000000314be75fae in __libc_message (do_abort=2, 
    fmt=0x314bf74af8 "*** glibc detected *** %s: %s: 0x%s ***\n")
    at ../sysdeps/unix/sysv/linux/libc_fatal.c:198
#3  0x000000314be7c2d6 in malloc_printerr (action=3, 
    str=0x314bf71aac "corrupted double-linked list", ptr=<optimized out>)
    at malloc.c:5021
#4  0x000000314be7c7a9 in malloc_consolidate (av=0x314c1af700)
    at malloc.c:4266
#5  0x000000314be7d3c9 in malloc_consolidate (av=0x314c1af700)
    at malloc.c:4227
#6  _int_free (av=0x314c1af700, p=<optimized out>, have_lock=0)
    at malloc.c:4158
#7  0x00007fcde48e426d in ?? ()
   from /home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so
#8  0x00007fcde48e551e in ?? ()
   from /home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so
#9  0x00007fcde494797d in ?? ()
   from /home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so
#10 0x00007fcde49479a5 in ?? ()
   from /home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so
#11 0x00007fcde3c6d0b9 in ?? ()
   from /home/nicolas/Projects/OpenCL/AMD-APP-SDK-v2.6-RC3-lnx64/lib/x86_64/libamdocl64.so
#12 0x000000314be39931 in __run_exit_handlers (status=0, listp=0x314c1af668, 
    run_list_atexit=true) at exit.c:78
#13 0x000000314be399b5 in __GI_exit (status=<optimized out>) at exit.c:100
#14 0x00000000004d21a8 in stg_exit ()
#15 0x00000000004d21cd in shutdownHaskellAndExit ()
#16 0x00000000004daab1 in real_main ()
#17 0x00000000004daafe in hs_main ()
#18 0x000000314be2169d in __libc_start_main (main=0x404030 <main>, argc=1, 
    ubp_av=0x7fff96abc628, init=<optimized out>, fini=<optimized out>, 
    rtld_fini=<optimized out>, stack_end=0x7fff96abc618) at libc-start.c:226
#19 0x0000000000406b45 in _start ()

Once in a while execution crashes with this backtrace:

#0  0x000000314be7c65f in malloc_consolidate (av=0x314c1af700)
    at malloc.c:4258
#1  0x000000314be7d3c9 in malloc_consolidate (av=0x314c1af700)
    at malloc.c:4227
#2  _int_free (av=0x314c1af700, p=<optimized out>, have_lock=0)
    at malloc.c:4158
(rest as above)

I'm not sure, but this looks like a double-free to me.

This might obviously be a bug in the AMD SDK, not sure...

Using ghc-7.0.4-31.3.fc16.x86_64, AMD-APP-SDK-v2.6-RC3-lnx64 and (Haskell) OpenCL 1.0.2.4.

Memory use is too large!

I make little modification to your example code example0.hs.

I change the vector size to 10000000 or larger. Because the length is long so the list will not be efficient enough. I use the Vector to store my data. Then I monitored the memory use of my program. It seems that everything is fine until the line mem_in <- V.unsafeWith vec (\x -> clCreateBuffer context [CL_MEM_READ_ONLY, CL_MEM_COPY_HOST_PTR] (vecSize, castPtr x))

Before that, the memory usage is reasonable and it is about 1MB , however, with this line and following lines added, the program uses too much memory which is up to 137MB!

I wonder if this is any bug in this.

import Control.Parallel.OpenCL
import Foreign( castPtr, nullPtr, sizeOf )
import Foreign.C.Types( CFloat )
import Foreign.Marshal.Array( newArray, peekArray )
import qualified Data.Vector.Storable as V
programSource :: String
programSource = "__kernel void duparray(__global float *in, __global float *out ){\n  int id = get_global_id(0);\n  out[id] = 2*in[id];\n}"

main :: IO ()
main = do
  -- Initialize OpenCL
  (platform:_) <- clGetPlatformIDs
  (dev:_) <- clGetDeviceIDs platform CL_DEVICE_TYPE_GPU
  context <- clCreateContext [CL_CONTEXT_PLATFORM platform] [dev] print
  q <- clCreateCommandQueue context dev []

  -- Initialize Kernel
  program <- clCreateProgramWithSource context programSource
  clBuildProgram program [dev] ""
  kernel <- clCreateKernel program "duparray"

  -- Initialize parameters
  let vec = V.enumFromTo 1 (10000000::CFloat)
  let elemSize = sizeOf (0 :: CFloat)
      len = 10000000
      vecSize = elemSize * len

  mem_in <- V.unsafeWith vec (\x -> clCreateBuffer context [CL_MEM_READ_ONLY, CL_MEM_COPY_HOST_PTR] (vecSize, castPtr x))  
--  mem_out <- clCreateBuffer context [CL_MEM_WRITE_ONLY] (vecSize, nullPtr)

--  clSetKernelArgSto kernel 0 mem_in
--  clSetKernelArgSto kernel 1 mem_out

  -- Execute Kernel
--  eventExec <- clEnqueueNDRangeKernel q kernel [len] [1] []
--  
--  -- Get Result
--  eventRead <- V.unsafeWith vec (\x -> clEnqueueReadBuffer q mem_out True 0 vecSize (castPtr x) [eventExec])
--  
--  result <- V.unsafeWith vec (\x-> peekArray (len) x)
--  putStrLn $ "Result array = " 
  return ()

API (clBuildProgram) is not referentially transparent

Your code from example01.hs has

  -- Initialize Kernel
  program <- clCreateProgramWithSource context programSource
  clBuildProgram program [dev] ""
  kernel <- clCreateKernel program "duparray"

Here clBuildProgram changes the program (from source to binary). This breaks referential transparency.

A more idiomatic approach would be to have clBuildProgram return the new program.

   -- Initialize Kernel
  program <- clCreateProgramWithSource context programSource
  program' <- clBuildProgram program [dev] ""
  kernel <- clCreateKernel program' "duparray"

This way you can also leverage the type system to prevent uncompiled programs being passed to clCreateKernel.