0

cudaカーネル関数内でforループに取り組みました。データを行と列に分割します。「fillFirstCells_kernel」が正しい結果を提供できない理由がわかりません。「fillFirstCells_kernel1」関数は、期待される結果を提供します。

コメント。

以下はコードです：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"

#include <stdio.h>
#include <stdlib.h>
#include <string.h>

#define ROWS    10
#define COLS    4

__global__ void fillData_kernel(int len,
                                unsigned int* data)
{
    int offset = blockIdx.x + blockDim.x * threadIdx.x;

    if (offset < len)
    {
        data[offset] = offset;

        offset += blockDim.x * gridDim.x;
    }
}

__global__ void fillFirstCells_kernel(unsigned int *data,
                                      unsigned int *result)
{
    int offset = blockIdx.x + blockDim.x * threadIdx.x;

    while (offset < ROWS)
    {
        for (int i=0; i<4; i++)
        {
            result[offset] += data[offset*COLS+i];      
            //result[offset] += data[offset*COLS];
            //atomicAdd(&result[offset], 1);
        }

        offset += blockDim.x * gridDim.x;
    }
}

__global__ void fillFirstCells_kernel1(unsigned int *data,
                                      unsigned int *result)
{
    int offset = blockIdx.x + blockDim.x * threadIdx.x;

    if (offset < ROWS)
    {
        result[offset] = data[offset*COLS] + 
                         data[offset*COLS+1] +
                         data[offset*COLS+2] +
                         data[offset*COLS+3];

        //offset += blockDim.x * gridDim.x;
    }
}

void displayIntOutput(unsigned int* data)
{
    for (unsigned int i=0; i<ROWS; i++)
    {
        for (unsigned int j=0; j<COLS; j++)
        {
            printf("%08x ", data[i*COLS+j]);
        }
        printf("\n");
    }

    printf("\n");
}


void main()
{
    unsigned int *h_data, *h_filled_data;
    unsigned int *d_filled_data;

    int size = ROWS * COLS * sizeof(unsigned int);

    h_data = (unsigned int*)malloc(size);
    h_filled_data = (unsigned int*)malloc(size);

    cudaMalloc((void**)&d_filled_data, size);
    cudaMemset(d_filled_data, 0, size);

    dim3 threads(8, 1);
    dim3 blocks(ceil((float)size/(float)threads.x), 1);

    fillData_kernel<<<blocks,threads>>>(size, d_filled_data);

    cudaMemcpy(h_filled_data, d_filled_data, size, cudaMemcpyDeviceToHost);

    displayIntOutput(h_filled_data);


    unsigned int *h_first_item_in_col;
    unsigned int *d_first_item_in_col;

    h_first_item_in_col = (unsigned int*)malloc(ROWS*sizeof(int));
    memset(h_first_item_in_col, 0, ROWS*sizeof(int));

    cudaMalloc((void**)&d_first_item_in_col, ROWS*sizeof(int));
    cudaMemset(d_first_item_in_col, 0, ROWS*sizeof(int));

    dim3 threads_first(8, 1);
    dim3 blocks_first(ceil((float)(ROWS*sizeof(int))/(float)threads_first.x), 1);

    fillFirstCells_kernel<<<blocks, threads>>>(d_filled_data, d_first_item_in_col);

    cudaMemcpy(h_first_item_in_col, d_first_item_in_col, ROWS*sizeof(int), cudaMemcpyDeviceToHost);

    for (int i=0; i<ROWS; i++)
        printf("%d. %08x\n", (i+1), h_first_item_in_col[i]);

    cudaFree(d_filled_data);
    cudaFree(d_first_item_in_col);

    free(h_data);
    free(h_filled_data);
    free(h_first_item_in_col);

    system("pause");
}

カーネル「fillFirstCells_kernel」の出力:

00000000 00000001 00000002 00000003 00000004 00000005 00000006 00000007 00000008 00000009 0000000a 0000000b 0000000c 0000000d 0000000e 0000000f 00000010 00000011 00000012 00000013 00000014 00000015 00000016 00000017 00000018 00000019 0000001a 0000001b 0000001c 0000001d 0000001e 0000001f 00000020 00000021 00000022 00000023 00000024 00000025 00000026 00000027

  1. 00000006
  2. 00000016
  3. 00000026
  4. 00000036
  5. 00000046
  6. 00000056
  7. 00000066
  8. 00000076
  9. 0000010c
  10. 0000012c

カーネル「fillFirstCells_kernel」の出力:

00000000 00000001 00000002 00000003 00000004 00000005 00000006 00000007 00000008 00000009 0000000a 0000000b 0000000c 0000000d 0000000e 0000000f 00000010 00000011 00000012 00000013 00000014 00000015 00000016 00000017 00000018 00000019 0000001a 0000001b 0000001c 0000001d 0000001e 0000001f 00000020 00000021 00000022 00000023 00000024 00000025 00000026 00000027

  1. 00000006
  2. 00000016
  3. 00000026
  4. 00000036
  5. 00000046
  6. 00000056
  7. 00000066
  8. 00000076
  9. 00000086
  10. 00000096
4

1 に答える 1

3

あなたが犯した間違いがいくつかあります。 

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

次のようにする必要があります。

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

カーネルを起動するときは、バイトではなく要素ごとに 1 つのスレッドを起動する必要があります。

dim3 blocks(ceil((float)size/(float)threads.x), 1);

する必要があります

dim3 blocks(ceil((float)ROWS*COLS/(float)threads.x), 1);

カーネル関数の長さの引数が間違っています:

fillData_kernel<<<blocks,threads>>>(size, d_filled_data);

次のようにする必要があります。

fillData_kernel<<<blocks,threads>>>(ROWS*COLS, d_filled_data);

それだけだと思います。

于 2013-02-08T13:19:43.213 に答える