How to pass a list of strings to an opencl kernel using pyopencl?
Question: How to pass list of strings to an opencl kernel the right way?

I tried this way using buffers (see following code), but I failed.

OpenCL (struct.cl):

    
    
    typedef struct{
              uchar uc[40];
    } my_struct9; 
    
    inline void try_this7_now(__global const uchar * IN_DATA ,
                              const uint IN_len_DATA ,
                              __global uchar * OUT_DATA){
        for (unsigned int i=0; i<IN_len_DATA ; i++)  OUT_DATA[i] = IN_DATA[i];
    }
    
    __kernel void try_this7(__global const my_struct9 * pS_IN_DATA ,
                            const uint IN_len ,
                            __global my_struct9 * pS_OUT){
    
        uint idx = get_global_id(0);
    for (unsigned int i=0; i<idx; i++) try_this7_now(pS_IN_DATA[i].uc, IN_len, pS_OUT[i].uc);
      }
    

Python (opencl_struct.py):

    
    
    # -*- coding: utf-8 -*- 
    
    import pyopencl as cl
    import pyopencl.array as cl_array
    import numpy
    
    ctx = cl.create_some_context()
    queue = cl.CommandQueue(ctx)
    # --------------------------------------------------------
    LIMIT = 40
    mf = cl.mem_flags
    
    import ctypes,sys,struct
    """
    typedef struct{
              uchar uc[40];
    } my_struct9; 
    """
    INlist = []
    INlist.append("That is VERY cool!")
    INlist.append("It is a list!")
    INlist.append("A big one!")
    #INlist.append("But it failes to output. :-(")  # PLAY WITH THOSE
    INlist.append("WTF is THAT?") # PLAY WITH THOSE
    print "INlist : "+str(INlist)
    print "largest string "+str( max( len(INlist[iL]) for iL in range(len(INlist)) ) )
    strLIMIT=str(LIMIT)
    s7 = struct.Struct(  (str(strLIMIT+'s') *len(INlist)) )
    IN_host_buffer = ctypes.create_string_buffer(s7.size)
    s7.pack_into(IN_host_buffer, 0, *INlist)
    IN_dev_buffer = cl.Buffer(ctx, mf.READ_ONLY | mf.COPY_HOST_PTR, hostbuf=IN_host_buffer)
    
    OUT_host_buffer = ctypes.create_string_buffer(s7.size)
    
    OUT_dev_buffer = cl.Buffer(ctx, mf.WRITE_ONLY, len(OUT_host_buffer))
    print "> len(OUT_host_buffer) "+str(len(OUT_host_buffer))
    
    # ========================================================================================
    f = open("struct.cl", 'r')
    fstr = "".join(f.readlines())
    prg = cl.Program(ctx, fstr).build()
    
    #cl.enqueue_copy(queue, IN_dev_buffer, IN_host_buffer, is_blocking=True) # copy data to device
    cl.enqueue_write_buffer(queue, IN_dev_buffer, IN_host_buffer).wait()
    
    prg.try_this7(queue, (1,), None, IN_dev_buffer, numpy.uint32(LIMIT), OUT_dev_buffer)
    # ========================================================================================
    cl.enqueue_copy(queue, OUT_host_buffer, OUT_dev_buffer).wait()
    
    SSS = s7.unpack_from(OUT_host_buffer,0)
    
    # unpack here OUT_host_buffer
    print "(GPU) output : "+str( SSS )+" "
    
    for s in range(len(SSS)):
     print ">>> (GPU) output : "+str( SSS[s] )
    

I ran the program first time with "but it failes to output" as 4th list
element. Then I played around by increasing and decreasing elements of the
list. Finally, there appeared this problem: **The output of the program is
supposed to be** (short version)

> > > > (GPU) output : That is VERY cool!
>>>>

>>>> (GPU) output : It is a list!

>>>>

>>>> (GPU) output : A big one!

>>>>

>>>> (GPU) output : WTF is THAT?

But it is:

> python opencl_struct.py
>
> INlist : ['That is VERY cool!', 'It is a list!', 'A big one!', 'WTF is
> THAT?']
>
> largest string 18
>

>> len(OUT_host_buffer) 160 (GPU) output : ('That is VERY
cool!\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00',
'It is a
list!\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00',
'A big
one!\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00',
'But it failes to output.
:-(\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00\x00')

>>

>>> > (GPU) output : That is VERY cool!

>>>>

>>>> (GPU) output : It is a list!

>>>>

>>>> (GPU) output : A big one!

>>>>

>>>> (GPU) output : But it failes to output. :-(

As you can see, the the 4th list element differes.

So, maybe my approach is wrong or there is a bug in pyopencl or somewhere
else.

I am using a NVidia 9400 GPU.

Rambo


Answer: You code seems to me very complicated. And some part are not very clear to me.
For instance, I don't see why you create only one work item:

    
    
    prg.try_this7(queue, (1,), None,...)
    

Which force you to loop through your strings (in the kernel) instead of using
the available parallelism. Anyhow, if I well understand, you want to send some
strings to the GPU copy them in another buffer, get them back in the host side
and display them.

If it's the case here is a version using only numpy and of course pyopencl:

    
    
    import numpy as np
    import pyopencl as cl
    
    
    ctx = cl.create_some_context()
    queue = cl.CommandQueue(ctx)
    #The kernel uses one workitem per char transfert
    prog_str = """kernel void foo(global char *in, global char *out, int size){
                      int idx = get_global_id(0);
                      if (idx < size){
                          out[idx] = in[idx];
                      }
               }"""
    prog = cl.Program(ctx, prog_str).build()
    #Note that the type of the array of strings is '|S40' for the length
    #of third element is 40, the shape is 3 and the nbytes is 120 (3 * 40)
    original_str = np.array(('this is an average string', 
                             'and another one', 
                             "let's push even more with a third string"))
    mf = cl.mem_flags
    in_buf = cl.Buffer(ctx, mf.READ_ONLY | mf.COPY_HOST_PTR, hostbuf=original_str)
    out_buf = cl.Buffer(ctx, mf.WRITE_ONLY, size=str_size)
    copied_str = np.zeros_like(original_str)
    #here launch the kernel with str_size number of workitems in this case 120
    #this mean that some of the workitems won't process any meaningful char 
    #(not all string have a lenght of 40) but it's no biggie
    prog.foo(queue, (str_size,), None, in_buf, out_buf, np.int32(str_size))
    cl.enqueue_copy(queue, copied_str, out_buf).wait()
    print copied_str
    

And the displayed result:

    
    
    ['this is an average string' 'and another one'
     "let's push even more with a third string"]
    

