✨🚧 Apply the enhancement of interopPlugin to samplePlugin

Author: davidAlgis

Plugin/SampleBasic/include/action_sample_texture.h

@@ -17,7 +17,6 @@ class ActionSampleTexture : public Action
 
     private:
     Texture *_texture;
-    cudaSurfaceObject_t _surf;
 };
 } // namespace SampleBasic
 

Plugin/SampleBasic/include/action_sample_texture_array.h

@@ -15,7 +15,6 @@ namespace SampleBasic {
 
 	private:
 		Texture* _texture;
-		cudaSurfaceObject_t* _surf;
 	};
 } // namespace SampleBasic
 

Plugin/SampleBasic/include/sample_kernels.cuh

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

Plugin/SampleBasic/src/action_sample_texture.cpp

@@ -10,30 +10,28 @@ ActionSampleTexture::ActionSampleTexture(void *texturePtr, int width,
     : Action()
 {
     _texture = CreateTextureInterop(texturePtr, width, height, 1);
-    _surf = 0;
 }
 
 inline int ActionSampleTexture::Start()
 {
     _texture->registerTextureInCUDA();
-    _surf = _texture->mapTextureToSurfaceObject();
+    _texture->mapTextureToSurfaceObject();
     return 0;
 }
 
 int ActionSampleTexture::Update()
 {
-    // _texture->copyUnityTextureToAPITexture();
     kernelCallerWriteTexture(_texture->getDimGrid(), _texture->getDimBlock(),
-                             _surf, GetTime(), _texture->getWidth(),
+                             _texture->getSurfaceObject(), GetTime(), _texture->getWidth(),
                              _texture->getHeight());
     // cudaDeviceSynchronize();
     return 0;
 }
 
 inline int ActionSampleTexture::OnDestroy()
 {
-    _texture->unMapTextureToSurfaceObject(_surf);
-    _texture->unRegisterTextureInCUDA();
+    _texture->unmapTextureToSurfaceObject();
+    _texture->unregisterTextureInCUDA();
     return 0;
 }
 

Plugin/SampleBasic/src/action_sample_texture_array.cpp

@@ -10,40 +10,29 @@ ActionSampleTextureArray::ActionSampleTextureArray(void *texturePtr, int width,
     : Action()
 {
     _texture = CreateTextureInterop(texturePtr, width, height, depth);
-    _surf = new cudaSurfaceObject_t[depth];
-    for(int i=0; i<depth; i++)
-    {
-        _surf[i] = 0;
-    }
 }
 
 inline int ActionSampleTextureArray::Start()
 {
     _texture->registerTextureInCUDA();
-    _surf = _texture->mapTextureArrayToSurfaceObject();
+    _texture->mapTextureToSurfaceObject();
     return 0;
 }
 
 int ActionSampleTextureArray::Update()
 {
     kernelCallerWriteTextureArray(
         _texture->getDimGrid(), _texture->getDimBlock(),
-                             _surf[0], GetTime(), _texture->getWidth(),
-                             _texture->getHeight(), _texture->getDepth());
-    kernelCallerWriteTextureArray(
-        _texture->getDimGrid(), _texture->getDimBlock(),
-                             _surf[1], 2*GetTime(), _texture->getWidth(),
+                             _texture->getSurfaceObjectArray(), GetTime(), _texture->getWidth(),
                              _texture->getHeight(), _texture->getDepth());
-
     cudaDeviceSynchronize();
     return 0;
 }
 
 inline int ActionSampleTextureArray::OnDestroy()
 {
-    _texture->unMapTextureToSurfaceObject(_surf[0]);
-    _texture->unRegisterTextureInCUDA();
-    delete(_surf);
+    _texture->unmapTextureToSurfaceObject();
+    _texture->unregisterTextureInCUDA();
     return 0;
 }
 

Plugin/SampleBasic/src/action_sample_vertex_buffer.cpp

@@ -33,7 +33,7 @@ namespace SampleBasic {
 
 	inline int ActionSampleVertexBuffer::OnDestroy()
 	{
-		_vertexBuffer->unRegisterBufferInCUDA();
+		_vertexBuffer->unregisterBufferInCUDA();
 		return 0;
 	}
 

Plugin/SampleBasic/src/sample_kernels.cu

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