CUDA程序

使用共享内存之后CUDA程序给出的结果有错,而且每次都不同的结果,
这是什么原因导致的?
谢谢

LZ您好:

请您提供您代码和出错的具体信息,否则本帖将做水帖处理。

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之后就运行,给出的结果不一样。

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之间的执行顺序也不被保证,所以这个是不确定的行为。您每次执行得到不同的结果原因在此。

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

祝您好运~

谢谢指点

不客气的,虽然最近论坛不太稳定,但是依然欢迎您常来~