🚧 First version that work for openGL and DX11. But texture array has to be treated slice by slice...

Author: davidAlgis

Plugin/SampleBasic/include/sample_kernels.cuh

@@ -11,7 +11,7 @@ void kernelCallerWriteTexture(const dim3 dimGrid, const dim3 dimBlock,
                               const float t, const int width, const int height);
 
 void kernelCallerWriteTextureArray(const dim3 dimGrid, const dim3 dimBlock,
-                                   cudaSurfaceObject_t* inputSurfaceObj,
+                                   cudaSurfaceObject_t* surfObjArray,
                                    const float time, const int width,
                                    const int height, const int depth);
 

Plugin/SampleBasic/src/action_sample_texture.cpp

@@ -22,9 +22,9 @@ inline int ActionSampleTexture::Start()
 int ActionSampleTexture::Update()
 {
     kernelCallerWriteTexture(_texture->getDimGrid(), _texture->getDimBlock(),
-                             _texture->getSurfaceObject(), GetTime(), _texture->getWidth(),
-                             _texture->getHeight());
-    // cudaDeviceSynchronize();
+                             _texture->getSurfaceObject(), GetTime(),
+                             _texture->getWidth(), _texture->getHeight());
+    cudaDeviceSynchronize();
     return 0;
 }
 

Plugin/SampleBasic/src/action_sample_texture_array.cpp

@@ -23,8 +23,8 @@ int ActionSampleTextureArray::Update()
 {
     kernelCallerWriteTextureArray(
         _texture->getDimGrid(), _texture->getDimBlock(),
-                             _texture->getSurfaceObjectArray(), GetTime(), _texture->getWidth(),
-                             _texture->getHeight(), _texture->getDepth());
+        _texture->getSurfaceObjectArray(), GetTime(), _texture->getWidth(),
+        _texture->getHeight(), _texture->getDepth());
     cudaDeviceSynchronize();
     return 0;
 }

Plugin/SampleBasic/src/action_sample_vertex_buffer.cpp

@@ -28,6 +28,8 @@ namespace SampleBasic {
 
 
 		_vertexBuffer->unmapResources();
+		delete(v);
+
 		return 0;
 	}
 

Plugin/SampleBasic/src/sample_kernels.cu

@@ -15,19 +15,19 @@ __global__ void writeTex(cudaSurfaceObject_t surf, int width, int height,
     }
 }
 
-__global__ void writeTexArray(cudaSurfaceObject_t* surf, int width, int height,
+__global__ void writeTexArray(cudaSurfaceObject_t* surfObjArray, int width, int height,
                               int depth, float time)
 {
     const unsigned int x = blockIdx.x * blockDim.x + threadIdx.x;
     const unsigned int y = blockIdx.y * blockDim.y + threadIdx.y;
     const unsigned int z = blockIdx.z * blockDim.z + threadIdx.z;
 
-    if (x < width && y < height)
+    if (x < width && y < height && z < depth)
     {
 
-        float4 t = make_float4(z % 2, abs(cos(time)), 0, 1.0f);
+        float4 t = make_float4(z % 2, abs((z+1)*cos(time)), 0, 1.0f);
 
-        surf2Dwrite(t, surf[z], sizeof(float4) * x, y);
+        surf2Dwrite(t, surfObjArray[z], sizeof(float4) * x, y);
     }
 }
 
@@ -52,11 +52,15 @@ void kernelCallerWriteTexture(const dim3 dimGrid, const dim3 dimBlock,
 }
 
 void kernelCallerWriteTextureArray(const dim3 dimGrid, const dim3 dimBlock,
-                                   cudaSurfaceObject_t* inputSurfaceObj,
+                                   cudaSurfaceObject_t* surfObjArray,
                                    const float time, const int width,
                                    const int height, const int depth)
 {
-    writeTexArray<<<dimGrid, dimBlock>>>(inputSurfaceObj, width, height, depth, time);
+
+    writeTex<<<dimGrid, dimBlock>>>(surfObjArray[0], width, height, time);
+
+    writeTex<<<dimGrid, dimBlock>>>(surfObjArray[1], width, height, 2*time);
+    // writeTexArray<<<dimGrid, dimBlock>>>(surfObjArray, width, height, depth, time);
 }
 
 void kernelCallerWriteBuffer(const dim3 dimGrid, const dim3 dimBlock,