用户
 找回密码
 立即注册
hunter 该用户已被删除
发表于 2013-8-20 07:03:22
69415
使用共享内存之后CUDA程序给出的结果有错,而且每次都不同的结果,
这是什么原因导致的?
谢谢

使用道具 举报 回复
发表于 2013-8-20 08:28:44
LZ您好:

请您提供您代码和出错的具体信息,否则本帖将做水帖处理。
使用道具 举报 回复 支持 反对
发表于 2013-8-21 05:47:44
__device__ float compute_d( float *GpuArray_d, int AttributeN_d, int z_d, int i_d_d)
{
   extern __shared__ float CurrentRow[];
    for(int zz = 0;  zz < AttributeN_d; zz++)
    {
        CurrentRow[zz] = GpuArray_d[z_d * AttributeN_d+ zz];

    }

    float tempdist_d = 0;
    for (int j = 0; j < AttributeN_d ; j++ ) {
         tempdist_d = tempdist_d + (powf(GpuArray_d[i_d_d*AttributeN_d+j] - CurrentRow[j],2.0));
          __syncthreads();

      }
    return tempdist_d;
}

__global__ void filterData_D(float *GpuArray, int Dataset_Dimension_d, int AttributeN_d, float *tempArrary_d, int i_d,clock_t *time_d)
{
  const int bid = blockIdx.x;
  const int tid = threadIdx.x;
  //if(tid == 0) time_d[bid] = clock();
  float tempdist = 0;
  for ( int z = bid * THREAD_NUM + tid;  z < Dataset_Dimension_d; z += BLOCK_NUM * THREAD_NUM)
  {
      tempdist = compute_d(GpuArray, AttributeN_d, z, i_d);
      tempArrary_d[z] = (sqrtf(tempdist));
      tempdist = 0;
  }
}
这是调用KERNEL部分
filterData_D <<< BLOCK_NUM, THREAD_NUM,49152,0>>>( gpudata, E_Dataset_Dimension, E_AttributeN, tempArrary_H, i, E_time );
AttributeN_d=168;
在compute_d里定义以及赋值CurrentRow[]之前正常,定义CurrentRow[]之后就运行,给出的结果不一样。
使用道具 举报 回复 支持 反对
发表于 2013-8-21 09:42:25
hunter 发表于 2013-8-21 05:47
__device__ float compute_d( float *GpuArray_d, int AttributeN_d, int z_d, int i_d_d)
{
   extern _ ...

LZ您好:

根据您3#提供的代码,目测发现了明显的shared memory使用问题:

extern __shared__ float CurrentRow[];
    for(int zz = 0;  zz < AttributeN_d; zz++)
    {
        CurrentRow[zz] = GpuArray_d[z_d * AttributeN_d+ zz];

    }
从这里可以看到 CurrentRow[]是shared memory中的数组,在block内是共享的。
然后您在循环中对CurrentRow[0]~CurrentRow[AttributeN_d-1]进行赋值。

这里的问题在于AttributeN_d[]整个block只有一份,而您一个block中的每个线程都试图对其全部的元素进行赋值,而且赋值的内容是不同的,考虑到线程之间是乱序的,这在逻辑上本身就是一种不确定的行为。

具体到硬件行为,按照手册的说法,每个warp中只会有一个线程成功写入,并且不保证是哪个线程写入。再考虑到warp之间的执行顺序也不被保证,所以这个是不确定的行为。您每次执行得到不同的结果原因在此。

因此,请您重新考虑算法逻辑,并给出新的实现。

祝您好运~
使用道具 举报 回复 支持 反对
发表于 2013-8-21 20:49:08
谢谢指点
使用道具 举报 回复 支持 反对
发表于 2013-8-21 20:56:37
hunter 发表于 2013-8-21 20:49
谢谢指点

不客气的,虽然最近论坛不太稳定,但是依然欢迎您常来~
使用道具 举报 回复 支持 反对
发新帖
您需要登录后才可以回帖 登录 | 立即注册