使用CUDA计算图像渐变时的条纹

我正在为图像去噪编写代码，并遇到了处理图像中条纹的奇怪问题。基本上，当我计算图像的X梯度时，出现水平条纹（或垂直于Y方向）Lena X gradient。整个算法工作正常，它看起来像我得到正确的答案（我与C程序进行比较），除了那些令人讨厌的条纹Lena result。使用CUDA计算图像渐变时的条纹

条纹之间的距离随着不同的块大小而变化。每次我运行程序时，我也有不同的条纹位置！这是与梯度计算相关的程序的一部分。我有一种感觉，我正在做一件非常愚蠢的事情:)谢谢！

#define BLKXSIZE 16 
#define BLKYSIZE 16 
#define idivup(a, b) (((a)%(b) != 0) ? (a)/(b)+1 : (a)/(b)) 
void Diff4th_GPU(float* A, float* B, int N, int M, int Z, float sigma, int iter, float tau, int       type) 
{ 
    float *Ad; 

dim3 dimBlock(BLKXSIZE,BLKYSIZE); 
dim3 dimGrid(idivup(N,BLKXSIZE), idivup(M,BLKYSIZE)); 

cudaMalloc((void**)&Ad,N*M*sizeof(float));   


cudaMemcpy(Ad,A,N*M*sizeof(float),cudaMemcpyHostToDevice); 

cudaCheckErrors("cc1"); 
int n = 1; 
while (n <= iter) { 
    Diff4th2D<<<dimGrid,dimBlock>>>(Ad, N, M, sigma, iter, tau, type); 
    n++; 
    cudaDeviceSynchronize(); 
    cudaCheckErrors("kernel");} 

cudaMemcpy(B,Ad,N*M*sizeof(float),cudaMemcpyDeviceToHost); 
cudaCheckErrors("cc2"); 
cudaFree(Ad); 
} 

    __global__ void Diff4th2D(float* A, int N, int M, float sigma, int iter, float tau, int type) 
{ 

float gradX, gradX_sq, gradY, gradY_sq, gradXX, gradYY, gradXY, sq_sum, xy_2, Lam, V_norm, V_orth, c, c_sq, lam_t; 


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

int index = j + i*N; 

if ((i < N) && (j < M)) 
    { 
    float gradX = 0, gradY = 0, gradXX = 0, gradYY = 0, gradXY = 0; 

    if ((i>1) && (i<N)) { 
     if ((j>1) && (j<M)){ 
    int indexN = (j)+(i-1)*(N); 
    if (indexN > ((N*M)-1)) indexN = (N*M)-1; 
    if (indexN < 0) indexN = 0; 
    int indexS = (j)+(i+1)*(N); 
    if (indexS > ((N*M)-1)) indexS = (N*M)-1; 
    if (indexS < 0) indexS = 0; 
    int indexW = (j-1)+(i)*(N); 
    if (indexW > ((N*M)-1)) indexW = (N*M)-1; 
    if (indexW < 0) indexW = 0; 
    int indexE = (j+1)+(i)*(N);   
    if (indexE > ((N*M)-1)) indexE = (N*M)-1; 
    if (indexE < 0) indexE = 0; 

    gradX = 0.5*(A[indexN]-A[indexS]); 
    A[index] = gradX; 
    } 
} 
} 
}

来源

2013-03-15 Daniel

在内核中有一个竞争条件，因为A的元素在使用之前可能会被覆盖，也可能不会被覆盖。

使用不同的数组进行输入和输出。

来源

2013-03-15 15:28:44 tera

非常感谢，它做到了！看起来应该重新考虑CPU编程习惯:) – Daniel 2013-03-15 15:40:49

使用CUDA计算图像渐变时的条纹

回答

相关问题