util/X86MAC64/cuda/samples/3_Imaging/imageDenoising/imageDenoising_copy_kernel.cuh

   1 /*
   2  * Copyright 1993-2013 NVIDIA Corporation.  All rights reserved.
   3  *
   4  * Please refer to the NVIDIA end user license agreement (EULA) associated
   5  * with this source code for terms and conditions that govern your use of
   6  * this software. Any use, reproduction, disclosure, or distribution of
   7  * this software and related documentation outside the terms of the EULA
   8  * is strictly prohibited.
   9  *
  10  */
  11
  12
  13
  14 __global__ void Copy(
  15     TColor *dst,
  16     int imageW,
  17     int imageH
  18 )
  19 {
  20     const int ix = blockDim.x * blockIdx.x + threadIdx.x;
  21     const int iy = blockDim.y * blockIdx.y + threadIdx.y;
  22     //Add half of a texel to always address exact texel centers
  23     const float x = (float)ix + 0.5f;
  24     const float y = (float)iy + 0.5f;
  25
  26     if (ix < imageW && iy < imageH)
  27     {
  28         float4 fresult = tex2D(texImage, x, y);
  29         dst[imageW * iy + ix] = make_color(fresult.x, fresult.y, fresult.z, 0);
  30     }
  31 }
  32
  33 extern "C" void
  34 cuda_Copy(TColor *d_dst, int imageW, int imageH)
  35 {
  36     dim3 threads(BLOCKDIM_X, BLOCKDIM_Y);
  37     dim3 grid(iDivUp(imageW, BLOCKDIM_X), iDivUp(imageH, BLOCKDIM_Y));
  38
  39     Copy<<<grid, threads>>>(d_dst, imageW, imageH);
  40 }