0

这是我第二次尝试 CUDA 程序和并行开发。

我试图找出点积两个数组的原因。

例子：

给定 

A = [1 2 3] 
B = [4 5 6]

C[0] = (1)(4) + (1)(5) + (1)(6)  
C[1] = (2)(4) + (2)(5) + (2)(6)  
C[2] = (3)(4) + (3)(5) + (3)(6)

我初始化 2 个数组 A 和 B ，随机填充一个范围内的元素，然后将每个元素A与每个 in相乘B，并将乘积的总和存储在标识为 的第三个数组中C。我已将数组 A、B 和 C 的大小设置为 100。

这给了我 10'000 次乘法，这些乘法是我使用 100 个块和 128 个线程（由于扭曲大小）并行化的。

这是我的内核函数：

__global__ void kernel(float *a, float *b, float *c, const int N) {
    if( threadIdx.x < N ) 
        c[blockIdx.x] += a[blockIdx.x] * b[threadIdx.x];
}

这是我的推理，因为必须累积聚合，C其中的索引与枢轴索引相同，A因此我可以重用blockidx.x并且“应该”正常工作；但事实并非如此。

我的怀疑是C当线程更改时索引被清除或不共享，但我真的不确定这就是我寻求建议的原因。

这是完整的代码，为了简短起见，我明确避免使用HANDLE_ERROR函数包装器

#include <stdio.h>
#include <cuda.h>
#include <time.h>

#define M 100

__global__ void kernel(float *a, float *b, float *c, const int N) {
    if(threadIdx.x < N) 
        c[blockIdx.x] += a[blockIdx.x] * b[threadIdx.x];
}

void init_array(float*, const int);
void fill_array(float*, const int, const float); 
void print_array(float*, const int, *char);

int main (void) {
    srand( time(NULL) );

    float a[M], b[M], c[M] = { 0.0 };
    float *dev_a, *dev_b, *dev_c;
    const int S = sizeof(float) * M;

    init_array(a, M);
    init_array(b, M);

    print_array(a, M, "a");
    print_array(b, M, "b");
    print_array(c, M, "c");

    cudaMalloc((void**)&dev_a, S);
    cudaMalloc((void**)&dev_b, S);
    cudaMalloc((void**)&dev_c, S);

    cudaMemcpy(dev_a, a, S, cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b, b, S, cudaMemcpyHostToDevice);
    cudaMemcpy(dev_c, c, S, cudaMemcpyHostToDevice);

    kernel<<<M, M + 28>>>(dev_a, dev_b, dev_c, M);

    cudaMemcpy(c, dev_c, S, cudaMemcpyDeviceToHost);

    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_c);

    print_array(c, M, "c");

    return 0;
}

void init_array(float *a, const int N) {
   int i;  
   for(i=0; i<N; i++)
       a[i] = rand() % M + 1;
}

void fill_array(float *a, const int N, const float v) {
   int i;  
   for(i=0; i<N; i++)
       a[i] = v;
}

void print_array(float *a, const int N, char *d) {
   int i;  
   for(i=0; i<N; i++)
       printf("\n%s[%d]: %f",d, i, a[i]);
}
4

3 回答 3

1

使用 cuBLAS 会简单得多：cublasSdot()例程执行您要查找的操作（即点积两个向量）。

这并不能帮助您学习如何编写并行代码或在 CUDA 中工作，但它会给您带来良好的性能，并且会针对不同的 GPU 进行优化。最佳实践是尽可能使用库，除非有充分的理由不这样做。

另一个答案指出，您需要使用原子或其他方法来避免竞争。一种更有效的方法是让每个线程计算部分结果，进行逐块归约（参见 SDK 中的示例），最后让每个块中的一个线程在全局内存中进行原子加法以累积结果来自不同的块。

于 2012-09-18T09:16:06.033 回答
1

将结果累加到 C 中时，您会遇到互斥。您不能让多个线程更新同一个数组索引。解决它的一种方法是使用原子指令，在您的情况下是 atomicAdd(..)。

这不起作用的原因是块 0 中的线程 0 和 1 更新了 C 数组中的相同位置。你得到一个比赛条件。

于 2012-09-18T09:03:11.177 回答
-2

您的代码将多个值写入全局内存中的同一位置，而无需同步和考虑。您可以使用某种关键部分来修复它。这是执行您想要的代码：

#include <stdio.h>
#include <cuda.h>
#include <time.h>

#define M 5

__global__ void kernel(float *a, float *b, float *c, const int N) {
    __shared__ int lock;
    lock=0u;
    if(threadIdx.x < N) {
        __syncthreads();
        float mult = a[blockIdx.x] * b[threadIdx.x];

        int leaveLoop = 0;
        while (leaveLoop==0) {
            if (atomicExch(&lock, 1u) == 0u) {
                //critical section
                c[blockIdx.x] += mult;
                leaveLoop = 1;
                atomicExch(&lock,0u);
            }
        }
        __syncthreads();
    }
}

void init_array(float *a, const int N) {
    int i;  
    for(i=0; i<N; i++)
        a[i] = rand() % M + 1;
}

void fill_array(float *a, const int N, const float v) {
    int i;  
    for(i=0; i<N; i++)
        a[i] = v;
}

void print_array(float *a, const int N, char *d) {
    int i;  
    for(i=0; i<N; i++)
        printf("\n%s[%d]: %f",d, i, a[i]);
}

int main (void) {
    srand( time(NULL) );

    float a[M], b[M], c[M] = { 0.0 };
    float *dev_a, *dev_b, *dev_c;
    const int S = sizeof(float) * M;

    init_array(a, M);
    init_array(b, M);

    print_array(a, M, "a");
    print_array(b, M, "b");
    print_array(c, M, "c");

    cudaMalloc((void**)&dev_a, S);
    cudaMalloc((void**)&dev_b, S);
    cudaMalloc((void**)&dev_c, S);

    cudaMemcpy(dev_a, a, S, cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b, b, S, cudaMemcpyHostToDevice);
    cudaMemcpy(dev_c, c, S, cudaMemcpyHostToDevice);

    kernel<<<M, M + 28>>>(dev_a, dev_b, dev_c, M);

    cudaMemcpy(c, dev_c, S, cudaMemcpyDeviceToHost);

    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_c);

    print_array(c, M, "c");

    return 0;
}

为了实现聚合数组乘法之和的目标，我建议您尽可能使用@Tom 提出的建议。我编写此代码是因为有时使用您的代码而不是外部库更容易。

于 2012-09-18T09:49:13.010 回答