CUDA Texture Texture Memory Sample Program

Original link

1 /*2 * Copyright Xu Hongzhi (northwest. School of Information Engineering). All rights reserved.3 * data:2012-4-204 */5 //6 //This program demonstrates the use of 1D and 2D texture memory7#include <stdio.h>8#include <cutil_inline.h>9#include <iostream>Ten using namespacestd; One  Atexture<float> texref1d;//1D Texture -texture<float,2> texref2d;//2D Texture -  the //1D Texture manipulation functions -__global__voidTEXTURE1D (float*DST,intWinth) - { -     intx = threadidx.x + blockidx.x *blockdim.x; +     inty = threadidx.y + blockidx.y *blockdim.y; -     intoffset = x + y * blockdim.x *griddim.x; +  A     if(X<w && y<h) atDst[offset] =Tex1dfetch (texref1d, offset); - } - //2D Texture manipulation functions -__global__voidTexture2d (float*DST,intWinth) - { -     intx = threadidx.x + blockidx.x *blockdim.x; in     inty = threadidx.y + blockidx.y *blockdim.y; -     intoffset = x + y * blockdim.x *griddim.x; to  +Dst[offset] =tex2d (texref2d, x, y); - } the intMainintargcChar**argv) * { $Cut_device_init (argc, argv);//Start CUDAPanax Notoginseng     ///1D Texture Memory -cout <<"1D Texture"<<Endl; the     float*HOST1D = (float*) Calloc (Ten,sizeof(float));//Memory Raw Data +     float*HOSTRET1D = (float*) Calloc (Ten,sizeof(float));//Memory Save return Data A  the     float*DEV1D, *DEVRET1D;//Video Memory Data +     inti; -cout <<"HOST1D:"<<Endl; $      for(i =0; I <Ten; ++i)//initializing memory raw data $     { -Host1d[i] = i *2; -cout <<"  "<< Host1d[i] <<" "; the     } -Cutilsafecall (Cudamalloc (void* *) &AMP;DEV1D,sizeof(float)*Ten));//Request Memory SpaceWuyiCutilsafecall (Cudamalloc (void* *) &AMP;DEVRET1D,sizeof(float)*Ten)); theCutilsafecall (cudamemcpy (dev1d, HOST1D,sizeof(float)*Ten, Cudamemcpyhosttodevice));//Copy the memory data into the video -Cutilsafecall (Cudabindtexture (NULL, texref1d, dev1d,sizeof(float)*Ten));//bind the memory data and textures Wu      -texture1d<<<Ten,1>>> (DEVRET1D,Ten,1);//run 1D texture manipulation functions About  $Cutilsafecall (cudamemcpy (HOSTRET1D, DEVRET1D,sizeof(float)*Ten, Cudamemcpydevicetohost));//Copy the memory data into RAM -     //Print Memory Data -cout << Endl <<"HOSTRET1D:"<<Endl; -      for(i =0; I <Ten; ++i) Acout <<"  "<< Hostret1d[i] <<" "; +  theCutilsafecall (Cudaunbindtexture (texref1d));//Unbind -Cutilsafecall (Cudafree (dev1d));//Free memory Space $ Cutilsafecall (Cudafree (DEVRET1D)); theFree (HOST1D);//free up memory space the Free (HOSTRET1D); the      the     ///2D Texture Memory -cout << Endl <<"2D Texture"<<Endl; in     intwidth =5, height =3; the     float*HOST2D = (float*) Calloc (width*height,sizeof(float));//Memory Raw Data the     float*HOSTRET2D = (float*) Calloc (width*height,sizeof(float));//Memory return Data About  theCudaarray *cuarray;//Cuda Array the     float*DEVRET2D;//Video Memory Data the     intRow, col; +cout <<"host2d:"<<Endl; -      for(row =0; Row < height; ++row)//initializing memory raw data the     {Bayi          for(col =0; Col < width; ++Col) the         { theHost2d[row*width + col] = row +Col; -cout <<"  "<< Host2d[row*width + col] <<" "; -         } thecout <<Endl; the     } theCudachannelformatdesc Channeldesc = cudacreatechanneldesc<float>(); theCutilsafecall (Cudamallocarray (&cuarray, &channeldesc, width, height));//Request Memory Space -Cutilsafecall (Cudamalloc (void* *) &AMP;DEVRET2D,sizeof(float) *width*height)); theCutilsafecall (Cudabindtexturetoarray (texref2d, Cuarray));//bind the memory data and textures theCutilsafecall (Cudamemcpytoarray (Cuarray,0,0, HOST2D,sizeof(float) (*width*height, Cudamemcpyhosttodevice));//copying memory data into a CUDA array the 94 dim3 Threads (width, height); thetexture2d<<<1, Threads>>> (devret2d, width, height);//run 2D texture manipulation functions the  theCutilsafecall (cudamemcpy (hostret2d, DEVRET2D,sizeof(float) (*width*height, cudamemcpydevicetohost));//Copy the memory data into RAM98     //Print Memory Data Aboutcout <<"hostret2d:"<<Endl; -      for(row =0; Row < height; ++row)101     {102          for(col =0; Col < width; ++Col)103cout <<"  "<< Hostret2d[row*width + col] <<" ";104cout <<Endl; the     }106 107Cutilsafecall (Cudaunbindtexture (texref2d));//Unbind108Cutilsafecall (Cudafreearray (Cuarray));//Free memory Space109 Cutilsafecall (Cudafree (DEVRET2D)); theFree (HOST2D);//free up memory space111 Free (HOSTRET2D); the 113Cut_exit (argc, argv);//Exit Cuda the}

CUDA Texture Texture Memory Sample Program