4

1D配列のCUDAにHAARウェーブレット変換を実装しようとしました。

アルゴリズム

入力配列に8つのインデックスがあります

この条件if(x_index>=o_width/2 || y_index>=o_height/2)では、0、2、4、6であるはずの4つのスレッドがあり、それぞれの入力で2つのインデックスを処理する予定です。

avg.EGを計算します。スレッドIDが「0」の場合、avgは(input [0] + input [1])/ 2であり、同時にinput[0]-となる差分を取得します。残りのスレッドの平均など。

ここで重要なのは、出力の配置です。インデックス0、2、4、6を使用すると、出力を正しいインデックスに配置する際に問題が発生したため、出力用に別のthread_idを作成しました。

私の平均は最初の4つのインデックス、つまり出力の0,1,2,3に配置する必要があり、o_thread_idは0,1,2,3である必要があります。同様に、4,5,6,7に差を配置するために、コードに示すように、0,1,2,3を「4」でインクリメントしました。

問題

私の出力はすべてゼロとして出てきます!!! 私が何を変えても、私はそれを得ています。

コード

__global__ void cal_haar(int input[],float output [],int i_widthstep,int o_widthstep,int o_width,int o_height)
{

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

    if(x_index>=o_width/2 || y_index>=o_height/2) return;

    int i_thread_id=y_index*i_widthstep+(2*x_index);
    int o_thread_id=y_index*o_widthstep+x_index;

    float avg=(input[i_thread_id]+input[i_thread_id+1])/2;
    float diff=input[i_thread_id]-avg;
    output[o_thread_id]=avg;
    output[o_thread_id+4]=diff;

}

void haar(int input[],float output [],int i_widthstep,int o_widthstep,int o_width,int o_height)
{

    int * d_input;
    float * d_output;

    cudaMalloc(&d_input,i_widthstep*o_height);
    cudaMalloc(&d_output,o_widthstep*o_height);

    cudaMemcpy(d_input,input,i_widthstep*o_height,cudaMemcpyHostToDevice);

    dim3 blocksize(16,16);
    dim3 gridsize;
    gridsize.x=(o_width+blocksize.x-1)/blocksize.x;
    gridsize.y=(o_height+blocksize.y-1)/blocksize.y;

    cal_haar<<<gridsize,blocksize>>>(d_input,d_output,i_widthstep,o_widthstep,o_width,o_height);


    cudaMemcpy(output,d_output,o_widthstep*o_height,cudaMemcpyDeviceToHost);

    cudaFree(d_input);
    cudaFree(d_output);

}

以下は私の主な機能です:-

void main()
{
    int in_arr[8]={1,2,3,4,5,6,7,8};
    float out_arr[8];
    int i_widthstep=8*sizeof(int);
    int o_widthstep=8*sizeof(float);
    haar(in_arr,out_arr,i_widthstep,o_widthstep,8,1);

    for(int c=0;c<=7;c++)
    {cout<<out_arr[c]<<endl;}
    cvWaitKey();

}

出力としてゼロが表示されるので、どこが間違っているのか教えていただけますか?ありがとうございました。

4

1 に答える 1

5

コードの問題は次の条件です。

if(x_index>=o_width/2 || y_index>=o_height/2) return;

が与えられるとo_height = 1o_height/2 = 0o_heightint、なので、ここでは切り捨てを伴う整数除算があります)、スレッドは操作を実行しません。必要なことを実現するには、ここで浮動小数点演算を実行するか、andを使用(o_height+1)/2(o_width+1)/2ます。「算術」丸めを使用して除算を実行します( x_index >= (8+1)/2 /*= 4*/ && y_index >= (1+1)/2 /*= 1*/ )

さらに、Y次元に複数のスレッドがある場合、アドレス指定に問題があります。それ以降、i_thread_id計算o_thread_idが正しくなくなります(_withstepサイズはバイト単位ですが、配列インデックスとして使用します)。

于 2012-05-23T20:35:45.653 に答える