0

編集: ソリューションで編集し、ネストされた for ループを実装しようとしている人々を支援するためにいくつかのコードを書きました。

そのため、cuda カーネルに次のネストされた for ループを実装しようとしています。

for (unsigned in=1;in<=N;in++){         
     for (unsigned ie=1;ie<=N-1;ie+=2){
         // do some stuff with x[in and/or ie] and z[in and/or ie]
}

x と z は、ホストからコピーされた単純な 1D 配列です。

さて、ネストされた for ループの実装に関しては、私が知る限り、これは比較的単純なはずです (以下の idx および idy インデックスで見られるように、誰かこれを確認できますか?)。私の他の質問は、パフォーマンスに影響を与えると予想されるカーネルで if ステートメントと % を使用せずに、つまり 2 ずつインクリメントするにはどうすればよいですか?

__global__ void myKernel(double x, double z, int N){

    // define indecies for storing counters
    int i = blockIdx.x * blockDim.x + threadIdx.x;
    int j = blockIdx.y * blockDim.y + threadIdx.y;

    // define counters
    int in = blockIdx.x * blockDim.x + (threadIdx.x + 1); // 1-based indexing
    int ie = (blockIdx.y * blockDim.y + (threadIdx.y)) * 2 + 1;

    z[i*N+j] = ie; // stores odd indicies [ie]

    // do some stuff with (in-ie), x[ie and/or in], z[ie and/or in]...
}

int main(){
    int N = 5; //arbitrary size of arrays
    int threadsPerBlock = 8; // semi-arbitrary, max 512 but depends on GPU device
    dim3 block(threadsPerBlock, threadsPerBlock/2 + 1); // launch N/2 + 1 threads in the second dimension since we are only interested in odd numbered indices.
    dim3 grid( (N-1+threadsPerBlock)/threadsPerBlock, (N-1+threadsPerBlock)/threadsPerBlock);

    std::vector<double> x;
    std::vector<double> z;
    z.resize(N*N);
    x.resize(N);

    //copy to device(d_x, d_z);

    //call the kernel
    myKernel<<<grid, block>>>(d_x, d_z, N);

    // retrieve from device

    return 0;
    }

ありがとう

4

1 に答える 1

0

それ以外の:

int ie = blockIdx.y * blockDim.y + (threadIdx.y + 1);

これを 2n + 1 にします。

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

そして明らかに、必要なのは N x N/2 スレッドだけです。

于 2013-09-09T08:16:18.690 に答える