OpenCL in float4 tezava

Imam tezavo z float4 podatkovnim tipom pri OpenCL. Rad bi zmanjsal stevilo operacij in branj iz globalnega pomnilnika (ce se kernel izvaja na graficni kartici).
V programu, ki se izvaja na gostitelju (CPU) sem rezerviral in napolnil s podatki tabelo velikosti 4*stevilo elementov (to uporabim namesto stirih locenih tabel za koordinate x, y, z in naboj c). To tabelo potem prenasam na GPU kot "float4 atoms".
Tip float4 naj bi bil velik 128b (4x32b float) in bi bil zato primeren da vanj shranim za posamezen atom koordinate x, y, z in naboj c. Torej podobno kot pri strukturah.
Tezava je ce v kernelu, ki se izvaja na GPU uporabim float4 kar pripelje do napake. Primer:

__kernel void vector_add(__global float4 *atoms,__global float *R, const int limit, __global float *dx, __global float *dy,__global float *dz){
 
  int i = get_global_id(0);

  int n,j;
  float rr, e, f, cg, tf, xx, yy, zz;

  //moja float4 spremenljivka, ki naj bi nosila vse 4 podatke o enem atomu
  __global float4 atom=(float4)(0.0f, 0.0f, 0.0f, 0.0f);

  //tu skusam iz tabele atoms[] prebrati podatke v  mojo float4 spremenljivko atom
  atom.x=atoms[i*4+0] ;  
  atom.y=atoms[i*4+1] ;  
  atom.z=atoms[i*4+2] ;  
  atom.w=atoms[i*4+3] ;

// naprej se s trenutnim atomom, ki ga imam shranjenega v "atom" racuna razdalje in sile med vsemi preostalimi 

Izhajal sem iz razlage na tej spletni strani: http://www.khronos.org/message_boards/v...

Ali zna kdo povedati, kje je tezava in kaj bi moral spremenit, da bi mi deloval float4?