stim/cuda/cuda_texture.cuh

#ifndef STIM_CUDA_TEXTURE_H
#define STIM_CUDA_TEXTURE_H
#include <assert.h>
#include <stim/cuda/cudatools/error.h>
#include <cuda.h>
#include <cuda_runtime.h>
#include <cublas_v2.h>
#include <stdio.h>
//#include <GL/glew.h>
#include <GL/glut.h>
#include <sstream>
#include <stim/visualization/colormap.h>
#include <stim/cuda/cudatools/devices.h>
#include <stim/cuda/cudatools/threads.h>
#include <stim/math/vector.h>
///A container for the texture based methods used by the spider class.
namespace stim
{
	namespace cuda
	{
		class cuda_texture
		{
		public:
			cudaArray* srcArray;
			cudaGraphicsResource_t resource;
			struct cudaResourceDesc resDesc;
			struct cudaTextureDesc texDesc;
			cudaTextureObject_t tObj;
			float *result;
			///basic constructor that creates the texture with default parameters.
			cuda_texture()
			{
			memset(&texDesc, 0, sizeof(texDesc));
				texDesc.addressMode[0] 		= cudaAddressModeWrap;
				texDesc.addressMode[1] 		= cudaAddressModeWrap;
				texDesc.filterMode		= cudaFilterModePoint;
				texDesc.readMode		= cudaReadModeElementType;
				texDesc.normalizedCoords	= 0;
			}
			///basic destructor
			~cuda_texture()
			{
				UnmapCudaTexture();
				if(result != NULL)
					cudaFree(result);
			}
			///Enable the nromalized texture coordinates.
			///@param bool, 1 for on, 0 for off
			void
			SetTextureCoordinates(bool val)
			{
				if(val)
					texDesc.normalizedCoords	=	1;
				else	
					texDesc.normalizedCoords	= 	0;
			}
			///sets the dimension dim to used the mode at the borders of the texture.
			///@param dim : 0-x, 1-y, 2-z
			///@param mode: cudaAddressModeWrap = 0,
			///		cudaAddressModeClamp = 1,
			///		cudaAddressMNodeMirror = 2,
			///		cudaAddressModeBorder = 3,
			void
			SetAddressMode(int dim, int mode)
			{
				switch(mode)
				{
					case 0:
						texDesc.addressMode[dim] = cudaAddressModeWrap;
						break;
					case 1:
						texDesc.addressMode[dim] = cudaAddressModeClamp;
						break;
					case 2:
						texDesc.addressMode[dim] = cudaAddressModeMirror;
						break;
					case 3:
						texDesc.addressMode[dim] = cudaAddressModeBorder;
						break;
					default:
						break;
				}
			}
//-------------------------------------------------------------------------//
//-------------------------------CUDA_MAPPING------------------------------//
//-------------------------------------------------------------------------//
//Methods for creating the cuda texture.
		///@param GLuint tex	 	 -- GLtexture (must be contained in a frame buffer object)
		///				    that holds that data that will be handed to cuda.
		///@param GLenum target		 -- either GL_TEXTURE_1D, GL_TEXTURE_2D or GL_TEXTURE_3D
		///				    map work with other gl texture types but untested.
		///Maps the gl texture in cuda memory, binds that data to a cuda array, and binds the cuda
		///array to a cuda texture.
		void
		MapCudaTexture(GLuint tex, GLenum target)
		{
			HANDLE_ERROR(
				cudaGraphicsGLRegisterImage(
					&resource,
					tex,
					target,
					cudaGraphicsMapFlagsReadOnly
//					cudaGraphicsRegisterFlagsNone	
				)
			);
			HANDLE_ERROR(
				cudaGraphicsMapResources(1, &resource)
			);
			
			HANDLE_ERROR(
				cudaGraphicsSubResourceGetMappedArray(&srcArray, resource, 0, 0)
			);
			
			memset(&resDesc, 0, sizeof(resDesc));
			resDesc.resType = cudaResourceTypeArray;
			resDesc.res.array.array = srcArray;
			HANDLE_ERROR(
				cudaCreateTextureObject(&tObj, &resDesc, &texDesc, NULL)
			);
		}
		///Unmaps the gl texture,  binds that data to a cuda array, and binds the cuda
		///array to a cuda texture.
		void
		UnmapCudaTexture()
		{
		//	HANDLE_ERROR(
		//		cudaGraphicsUnmapResources(1, &resource)
		//	);
		//	HANDLE_ERROR(
		//		cudaGraphicsUnregisterResource(resource)
		//	);	
		//	HANDLE_ERROR(
		//		cudaDestroyTextureObject(tObj)
		//	);
//			HANDLE_ERROR(
//				cudaFreeArray(srcArray)
//			);
		}
		///Allocate the auxiliary internal 1D float array
		void
		Alloc(int x)
		{
			cudaMalloc( (void**) &result, x*sizeof(float));
		}
//-------------------------------------------------------------------------//
//------------------------------GET/SET METHODS----------------------------//
//-------------------------------------------------------------------------//
///Returns the bound texture object.
		cudaTextureObject_t
		getTexture()
		{
			return tObj;
		}
		cudaArray*
		getArray()
		{
			return srcArray;
		}
		float*
		getAuxArray()
		{
			return result;
		}
	};
}
}
#endif