             delete testSplit;
         }
 
-        sphr_bvh_in_array   = new eavlConstArrayV2<float4>( (float4*)sphr_bvh_in_raw, sphr_bvh_in_size/4, sphr_bvh_in_tref);
-        sphr_bvh_lf_array   = new eavlConstArrayV2<float>( sphr_bvh_lf_raw, sphr_bvh_lf_size, sphr_bvh_lf_tref);
-        sphr_verts_array    = new eavlConstArrayV2<float4>( (float4*)sphr_verts_raw,numSpheres, sphr_verts_tref);
-        sphr_matIdx_array   = new eavlConstArrayV2<int>( sphr_matIdx_raw, numSpheres,  sphr_matIdx_tref );
-        sphr_scalars_array  = new eavlConstArrayV2<float>(sphr_scalars_raw, numSpheres, sphr_scalars_tref);
+        sphr_bvh_in_array   = new eavlConstArrayV2<float4>( (float4*)sphr_bvh_in_raw, sphr_bvh_in_size/4, sphr_bvh_in_tref, cpu);
+        sphr_bvh_lf_array   = new eavlConstArrayV2<float>( sphr_bvh_lf_raw, sphr_bvh_lf_size, sphr_bvh_lf_tref, cpu);
+        sphr_verts_array    = new eavlConstArrayV2<float4>( (float4*)sphr_verts_raw,numSpheres, sphr_verts_tref, cpu);
+        sphr_matIdx_array   = new eavlConstArrayV2<int>( sphr_matIdx_raw, numSpheres,  sphr_matIdx_tref, cpu );
+        sphr_scalars_array  = new eavlConstArrayV2<float>(sphr_scalars_raw, numSpheres, sphr_scalars_tref, cpu);
         /*no need for normals, trivial calculation */
         //INIT(eavlConstArray<int>, sphr_matIdx,numSpheres);
 

                                             IntMemsetFunctor(-2)),
                                             "init");
    eavlExecutor::Go();
    //eavlExecutor::AddOperation(new_eavlMapOp(eavlOpArgs(rayDirX,rayDirY,rayDirZ,rayOriginX,rayOriginY,rayOriginZ,hitIdx,primitiveTypeHit,minDistances),
    //                                             eavlOpArgs(dummy, depthBuffer, primitiveTypeHit),
    //                                             RayIntersectFunctor(tri_verts_array,tri_bvh_in_array,tri_bvh_lf_array, TRIANGLE )),
    //                                                                                                "intersect");
    //eavlExecutor::Go();

    eavlExecutor::AddOperation(new_eavlMapOp(eavlOpArgs(rayDirX,rayDirY,rayDirZ,rayOriginX,rayOriginY,rayOriginZ,hitIdx,primitiveTypeHit,minDistances),
                                             eavlOpArgs(dummy, depthBuffer, primitiveTypeHit),
                                             RayIntersectFunctor(sphr_verts_array,sphr_bvh_in_array,sphr_bvh_lf_array,SPHERE)),
                                                                                                    "intersect");
    eavlExecutor::Go();
    float maxDepth=0;
    float minDepth=INFINITE;


template<class T>
class eavlConstArrayV2
{
  private: 
    bool cpu;
  public:
    T *host;
    T *device;
    
  public:
    eavlConstArrayV2(T *from, int N, texture<T> &g_textureRef, bool CPU)
    {
        host = from;
        cpu = CPU;
        if(!CPU)
        {
#ifdef HAVE_CUDA
        
        
            int nbytes = N * sizeof(T);
            cudaMalloc((void**)&device, nbytes);
            CUDA_CHECK_ERROR();
            cudaMemcpy(device, &(host[0]),
                   nbytes, cudaMemcpyHostToDevice);
            CUDA_CHECK_ERROR();
#ifdef USE_TEXTURE_MEM
            cudaBindTexture(0, g_textureRef,device,nbytes);
            CUDA_CHECK_ERROR();
#endif

#endif
        }
    }
    ~eavlConstArrayV2()
    {
