Grayscale color image using CUDA parallel processing

I am trying to solve a problem in which I have to change a color image to a grayscale image. For this, I use the CUDA parallel approach.

The kernel code that I invoke on the GPU is as follows.

__global__
void rgba_to_greyscale(const uchar4* const rgbaImage,
                   unsigned char* const greyImage,
                   int numRows, int numCols)
{
    int absolute_image_position_x = blockIdx.x;  
    int absolute_image_position_y = blockIdx.y;

  if ( absolute_image_position_x >= numCols ||
   absolute_image_position_y >= numRows )
 {
     return;
 }
uchar4 rgba = rgbaImage[absolute_image_position_x + absolute_image_position_y];
float channelSum = .299f * rgba.x + .587f * rgba.y + .114f * rgba.z;
greyImage[absolute_image_position_x + absolute_image_position_y] = channelSum;

}

void your_rgba_to_greyscale(const uchar4 * const h_rgbaImage,
                            uchar4 * const d_rgbaImage,
                            unsigned char* const d_greyImage,
                            size_t numRows,
                            size_t numCols)
{
  //You must fill in the correct sizes for the blockSize and gridSize
  //currently only one block with one thread is being launched
  const dim3 blockSize(numCols/32, numCols/32 , 1);  //TODO
  const dim3 gridSize(numRows/12, numRows/12 , 1);  //TODO
  rgba_to_greyscale<<<gridSize, blockSize>>>(d_rgbaImage,
                                             d_greyImage,
                                             numRows,
                                             numCols);

  cudaDeviceSynchronize(); checkCudaErrors(cudaGetLastError());
}


I see a line of dots in the first row of the pixel.

the error I get is

      libdc1394 error: failed to initialize libdc1394
      Position 51 difference exceeds acceptable level 5 Reference: 255
      GPU: 0
my I / O images Can anyone help me with this ??? thanks in advance.

+5
source share
12

, , , , , . , .
, : -

 1. absolute_position_x =(blockIdx.x * blockDim.x) + threadIdx.x;
 2. absolute_position_y = (blockIdx.y * blockDim.y) + threadIdx.y;

-,

 1. const dim3 blockSize(24, 24, 1);
 2. const dim3 gridSize((numCols/16), (numRows/16) , 1);

numCols/16 * numCols/16
24 * 24

, 0.040576

@datenwolf: !

+5

, , . . :

__global__`
void rgba_to_greyscale(const uchar4* const rgbaImage,
               unsigned char* const greyImage,
               int numRows, int numCols)
{`

int pos_x = (blockIdx.x * blockDim.x) + threadIdx.x;
int pos_y = (blockIdx.y * blockDim.y) + threadIdx.y;
if(pos_x >= numCols || pos_y >= numRows)
    return;

uchar4 rgba = rgbaImage[pos_x + pos_y * numCols];
greyImage[pos_x + pos_y * numCols] = (.299f * rgba.x + .587f * rgba.y + .114f * rgba.z); 

}

, .

+5

. , . , pos_x pos_y numRows numCols. -, .

const dim3 blockSize(16, 16, 1);
const dim3 gridSize((numCols%16) ? numCols/16+1 : numCols/16,
(numRows%16) ? numRows/16+1 : numRows/16, 1);
+2

libdc1394: libdc1394

, CUDA. libdc1394 - , IEEE1394, FireWire, iLink ( DV, Apple iSight). , . NINO: Nonsens In Nonsens Out.

+1

x y . , ?

uchar4 rgba = rgbaImage[absolute_image_position_x + (absolute_image_position_y * numCols)];

, , , . , :)

+1

- .

:

  • uchar4 rgba = rgbaImage [absolute_image_position_x + absolute_image_position_y];
  • greyImage [absolute_image_position_x + absolute_image_position_y] = channelSum;

:

  • uchar4 rgba = rgbaImage [absolute_image_position_x + absolute_image_position_y * numCols];
  • greyImage [absolute_image_position_x + absolute_image_position_y * numCols] = channelSum;
+1
__global__
void rgba_to_greyscale(const uchar4* const rgbaImage,
                       unsigned char* const greyImage,
                       int numRows, int numCols)
{
    int rgba_x = blockIdx.x * blockDim.x + threadIdx.x;
    int rgba_y = blockIdx.y * blockDim.y + threadIdx.y;
    int pixel_pos = rgba_x+rgba_y*numCols;

    uchar4 rgba = rgbaImage[pixel_pos];
    unsigned char gray = (unsigned char)(0.299f * rgba.x + 0.587f * rgba.y + 0.114f * rgba.z);
    greyImage[pixel_pos] = gray;
}

void your_rgba_to_greyscale(const uchar4 * const h_rgbaImage, uchar4 * const d_rgbaImage,
                            unsigned char* const d_greyImage, size_t numRows, size_t numCols)
{
    //You must fill in the correct sizes for the blockSize and gridSize
    //currently only one block with one thread is being launched
    const dim3 blockSize(24, 24, 1);  //TODO
    const dim3 gridSize( numCols/24+1, numRows/24+1, 1);  //TODO
    rgba_to_greyscale<<<gridSize, blockSize>>>(d_rgbaImage, d_greyImage, numRows, numCols);

    cudaDeviceSynchronize(); checkCudaErrors(cudaGetLastError());
}
+1

:

  const dim3 blockSize(numCols/32, numCols/32 , 1);  //TODO
  const dim3 gridSize(numRows/12, numRows/12 , 1);  //TODO

!

 int absolute_image_position_x = blockIdx.x;  
 int absolute_image_position_y = blockIdx.y;

, absolute_image_position_x , absolute_image_position_y . / greyImage, . :)

0

int idx=blockDim.x*blockIdx.x+threadIdx.x;
int idy=blockDim.y*blockIdx.y+threadIdx.y;

uchar4 rgbcell=rgbaImage[idx*numCols+idy];

   greyImage[idx*numCols+idy]=0.299*rgbcell.x+0.587*rgbcell.y+0.114*rgbcell.z;


  }

  void your_rgba_to_greyscale(const uchar4 * const h_rgbaImage, uchar4 * const d_rgbaImage,
                        unsigned char* const d_greyImage, size_t numRows, size_t numCols)
 {
 //You must fill in the correct sizes for the blockSize and gridSize
 //currently only one block with one thread is being launched

int totalpixels=numRows*numCols;
int factors[]={2,4,8,16,24,32};
vector<int> numbers(factors,factors+sizeof(factors)/sizeof(int));
int factor=1;

   while(!numbers.empty())
  {
 if(totalpixels%numbers.back()==0)
 {
     factor=numbers.back();
     break;
 }
   else
   {
  numbers.pop_back();
   }
 }



 const dim3 blockSize(factor, factor, 1);  //TODO
 const dim3 gridSize(numRows/factor+1, numCols/factor+1,1);  //TODO
 rgba_to_greyscale<<<gridSize, blockSize>>>(d_rgbaImage, d_greyImage,    numRows, numCols);
0

In this case, the libdc1394 error is not related to firewire, etc. is the library that udacity uses to compare the image that your program creates for the reference image. And what it says is that the difference between your image and the reference image was exceeded by a certain threshold for this position, i.e. pixels.

0
source

1- int x =(blockIdx.x * blockDim.x) + threadIdx.x;

2- int y = (blockIdx.y * blockDim.y) + threadIdx.y;

And in grid and block size

1- const dim3 blockSize(32, 32, 1);

2- const dim3 gridSize((numCols/32+1), (numRows/32+1) , 1);

The code is executed in 0.036992 ms.

0
source
const dim3 blockSize(16, 16, 1);  //TODO
const dim3 gridSize( (numRows+15)/16, (numCols+15)/16, 1);  //TODO

int x = blockIdx.x * blockDim.x + threadIdx.x;  
int y = blockIdx.y * blockDim.y + threadIdx.y;

uchar4 rgba = rgbaImage[y*numRows + x];
float channelSum = .299f * rgba.x + .587f * rgba.y + .114f * rgba.z;
greyImage[y*numRows + x] = channelSum;
0
source

All Articles