24小时热门版块排行榜

返回列表

hsfy919

银虫 (初入文坛)

应助: 0 (幼儿园)
金币: 40.7
红花: 1
帖子: 12
在线: 14.5小时
虫号: 1145342
注册: 2010-11-12
性别: GG
专业: 动力学与控制

[求助] CUDA并行问题，全部金币悬赏

最近在研究CUDA做并行运算，刚开始接触，不是很明白共享内存（shared memory）的机理
问题是：我要利用CUDA函数做N1个数组相加，每个数组长度为N2，由于数组很多（大概上百万个）不能通过CPU直接传入GPU中，需要在GPU中通过计算每个线程得到一个数组，完成计算后所有线程的数组相加，得到长度为N2的一个数组，输出出来。可由于线程之间彼此不能通信，想用shared memory来解决求和，但结果不对（我编写的程序如下）。

这个程序正确的结果应该是C=[150,300,450,600,750]，但得到的结果却是C=[750,750,750,750,750]
希望大家看看，如果能解决问题，全部金币奉上

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <iostream>
#include <fstream>
using namespace std;

#define NumberOfBlock 128
#define ThreadPerBlock 256

__global__ void addKernel(int *c, int *p, int *a, int *b, int N)
{
int id = threadIdx.x + blockIdx.x*blockDim.x, cacheindex=threadIdx.x, i, j, ii;
double temp, temp2;
__shared__ double cache[ThreadPerBlock];
for(j=0;j<N;j++)  // 问题在于当j循环到1时，得到的值覆盖了数组C中j=0时的值，当j=2时覆盖了j=0和1的值，后面也是
{
i=id;
      temp=0;
while(i<N)
{
temp+=a*b[j];
i+=gridDim.x*blockDim.x;
}
//----对每个j规约求和--------------------------------------------------------------------
cache[cacheindex]=temp;
__syncthreads();
ii=blockDim.x/2;
while(ii!=0)
{
if(cacheindex<ii)
cache[cacheindex]+=cache[cacheindex+ii];
__syncthreads();
ii/=2;
}
if(cacheindex==0)
p[blockIdx.x]=cache[0];
__syncthreads();
temp2=0;
for(ii=0;ii<NumberOfBlock;ii++)
temp2+=p[ii];
__syncthreads();
c[j]=temp2;
//--------------------------------------------------------------------------------
}

}

int main()
{
const int size = 5;  //以5个元素举例，实际问题元素很多，过百万个
int a[size] = { 1, 2, 3, 4, 5 };
int b[size] = { 10, 20, 30, 40, 50 };
int c[size];
for(int i=0;i<size;i++)
c = 0;
int *dev_a = 0;
int *dev_b = 0;
int *dev_c = 0;
int *dev_p = 0;

int *p=new int [NumberOfBlock];
for(int i=0;i<NumberOfBlock;i++)
p=0;

cudaMalloc((void**)&dev_c, size * sizeof(int));
cudaMalloc((void**)&dev_p, NumberOfBlock * sizeof(int));
cudaMalloc((void**)&dev_a, size * sizeof(int));
cudaMalloc((void**)&dev_b, size * sizeof(int));
cudaMemcpy(dev_a, a, size * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(dev_b, b, size * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(dev_c, c, size * sizeof(int), cudaMemcpyHostToDevice);
cudaMemcpy(dev_p, p, NumberOfBlock * sizeof(int), cudaMemcpyHostToDevice);

addKernel<<<NumberOfBlock, ThreadPerBlock>>>(dev_c, dev_p, dev_a, dev_b, size);
cudaDeviceSynchronize();

cudaMemcpy(c, dev_c, size * sizeof(int), cudaMemcpyDeviceToHost);
for(int i=0;i<size;i++)
cout<<c<<endl;
cudaFree(dev_c);
cudaFree(dev_a);
cudaFree(dev_b);
getchar();
}