title: How to pass an array of vectors in pyOpenCL
body: <p>I'm moving a simulation into pyOpenCL and can't get my data access to work. I'm trying to supply a 1D array of vectors (well, actually several, but the example I've included just used one).</p>
  
  <p>Currently, several vectors are copied over just fine, but then the data is simply not what I supplied.</p>
  
  <p>I don't think I've posted here before, so apologies if any of the formatting/presentation is wrong. Also, I've just stripped out all the simulation code, so I realise this code is currently not actually doing anything, I just want to get the buffer passing correct.</p>
  
  <p>Thanks in advance.</p>
  
  <p>The kernel (kertest.py):</p>
  
  <pre><code>step1 = """
  #pragma OPENCL EXTENSION cl_amd_printf: enable
  #define X xdim
  #define Y ydim
  __kernel void k1(__global float3 *spins,
                   __local float3 *tile)
  {        
      ushort lid = 2 * get_local_id(0);
      ushort group = 2 * get_group_id(0);
      ushort num = get_num_groups(0);
      int lim = X*Y*3;
  
      for (ushort i = 0; i &lt; lim; i++)
          {
              if (lid == 0 &amp;&amp; group == 0)
              {
                  printf("%f :: %d\\n", spins[i].x, i);
              }
           }
  }"""
  </code></pre>
  
  <p>The code itself (gputest.py):</p>
  
  <pre><code>import kertest as k2D
  import numpy as np
  import pyopencl as cl
  
  class GPU_MC2DSim():
      def __init__(self, x, y):
          self.x = x
          self.y = y
  
          if x &gt;= y:
              self.xdim = int(self.x)
              self.ydim = int(self.y)
          else:
              self.xdim = int(self.y)
              self.ydim = int(self.x)
  
          if self.xdim % 2 != 0: self.xdim += 1
  
          if self.ydim % 2 != 0: self.ydim += 1
  
          self.M = np.ones((self.xdim*self.ydim, 3)).astype(np.float32)
          self.M[:, 1] += 1.0
          self.M[:, 2] += 2.0
  
          print self.M
  
      def simulate(self):
          ctx = cl.create_some_context()
          q = cl.CommandQueue(ctx)
          mf = cl.mem_flags
  
          #Pass buffer:
          M_buf = cl.Buffer(ctx, mf.READ_WRITE | mf.COPY_HOST_PTR, hostbuf = self.M)
  
          #Insert kernel parameters:
          params = {"xdim" : "%d" % (self.xdim),
                    "ydim" : "%d" % (self.ydim),
                    }
          for name in params:
              k2D.step1 = k2D.step1.replace(name, params[name])
  
          #Compile kernel:
          step1 = cl.Program(ctx, k2D.step1).build()
  
          locmem = cl.LocalMemory(self.xdim*4*4)
  
          step1.k1(q, ((self.xdim*self.ydim)/4,), (self.xdim/2,), M_buf, locmem).wait()
          return None
  
  xdim = 4
  ydim = 4
  sim = GPU_MC2DSim(xdim, ydim)
  sim.simulate()
  </code></pre>