当BlockSize 7和BlockSize 8时，库达菲代码的结果不同

Question

我正在使用Cudafy.NET，我在BlockSize上遇到了一些困难。它在某些情况下产生了不同的结果。不同之处就在这里：

//correct results when using this line
gpu.Launch(1, 7, "kernelfx_alldata", 10, devdata, devnmin, devnmax, devgmin, devgmax, devtest);

//incorrect results when using this line
gpu.Launch(1, 8, "kernelfx_alldata", 10, devdata, devnmin, devnmax, devgmin, devgmax, devtest);

对这个问题的详细解释：

我有10个项目要循环。GridSize是1。

案例1:当eGPUType.OpenCL = CudafyModes.Target和BlockSize为1,2,3,4,5,6和7时。结果是正确的。 案例2: eGPUType.OpenCL = CudafyModes.Target，BlockSize是8、9、10、11、.等。结果是不正确的。 判例3: eGPUType.Emulator = BlockSize = 1,2,3,4,5,6,7,8,9,10,11，……还有更多的。结果是正确的。

示例代码如下所示。初始化变量：

double[,] data;
double[] nmin, nmax, gmin, gmax;

void initializeVars()
{
    data = new double[10, 10];
    for (int i = 0; i < 10; i++)
        {
            data[i, 0] = 100 + i;
            data[i, 1] = 32 + i;
            data[i, 2] = 22 + i;
            data[i, 3] = -20 - i;
            data[i, 4] = 5522 + 10 * i;
            data[i, 5] = 40 + i;
            data[i, 6] = 14 - i;
            data[i, 7] = 12 + i;
            data[i, 8] = -10 + i;
            data[i, 9] = 10 + 10 * i;
        }
    nmin = new double[10];
    nmax= new double[10];
    gmin = new double[10];
    gmax = new double[10];
    for (int i = 0; i < 10; i++)
    {
        nmin[i] = -1;
        nmax[i] = 1;
        gmin[i] = i;
        gmax[i] = 11 * i*i+1;
    }
}

gpu发射代码：

private void button1_Click(object sender, EventArgs e)
{
    CudafyModes.Target = eGPUType.OpenCL;
    CudafyModes.DeviceId = 0;
    CudafyTranslator.Language = eLanguage.OpenCL;
    CudafyModule km = CudafyTranslator.Cudafy();
    Cudafy.Host.GPGPU gpu = Cudafy.Host.CudafyHost.GetDevice(CudafyModes.Target, CudafyModes.DeviceId);
    gpu.LoadModule(km);
    initializeVars();
    double[,] devdata = gpu.Allocate<double>(data); gpu.CopyToDevice(data, devdata);
    double[] devnmin = gpu.Allocate<double>(nmin); gpu.CopyToDevice(nmin, devnmin);
    double[] devnmax = gpu.Allocate<double>(nmax); gpu.CopyToDevice(nmax, devnmax);
    double[] devgmin = gpu.Allocate<double>(gmin); gpu.CopyToDevice(gmin, devgmin);
    double[] devgmax = gpu.Allocate<double>(gmax); gpu.CopyToDevice(gmax, devgmax);
    double[] test = new double[10];
    double[] devtest = gpu.Allocate<double>(test);
    gpu.Launch(1, 8, "kernelfx_alldata", 10, devdata, devnmin,
           devnmax, devgmin, devgmax,  devtest);
    gpu.CopyFromDevice(devtest, test);
    gpu.FreeAll();
}

库达菲核

[Cudafy]
public static void kernelfx_alldata(GThread thread, int N, double[,] data, double[] nmin, double[] nmax, double[] gmin, double[] gmax, double[] test)
{
    int tid = thread.threadIdx.x + thread.blockIdx.x * thread.blockDim.x;
    while (tid < N)
    {
        double[] tmp = thread.AllocateShared<double>("tmp", 10);
        tmp[0] = 1; 
        for (int i = 1; i < 10; i++)
        {
            tmp[i] = data[tid, i - 1];
        }
        for (int i = 1; i < 10; i++)
        {
            tmp[i] = (nmax[i - 1] - nmin[i - 1]) / (gmax[i - 1] - gmin[i - 1]) * (tmp[i] - gmin[i - 1]) + nmin[i - 1];
        }
        test[tid] = tmp[1];

        tid = tid + thread.blockDim.x * thread.gridDim.x;
    }
}

正确的结果(案例1和案例3)是： test=199.0 test1=201.0 test2=203.0 test3=205.0 test4=207.0 test5=209.0 test6=211.0 test7=213.0 test8=215.0 test9=217.0 不正确(案例2)的结果是： test=213.0 test1=213.0 test2=213.0 test3=213.0 test4=213.0 test5=213.0 test6=213.0 test7=213.0 test8=217.0 test9=217.0

当BlockSize小于8时，结果是正确的。但是当BlockSize大于8时，结果是不正确的。为了有效地使用gpu，blockSize必须大于8。

这段代码有什么问题？

最诚挚的问候..。

Answer 1

将tmp声明为2d数组，第一列是threadId，解决了这个问题。工作守则如下：

[Cudafy]
public static void kernelfx_alldata(GThread thread, int N, double[,] data, double[] nmin,
                                double[] nmax, double[] gmin, double[] gmax, double[] test)
{
    int tid = thread.threadIdx.x + thread.blockIdx.x * thread.blockDim.x;
    double[,] tmp = thread.AllocateShared<double>("tmp", 10, 10);
    while (tid < N)
    {

        tmp[tid, 0] = 1;
        for (int i = 1; i < 10; i++)
        {
            tmp[tid, i] = data[tid, i - 1];
        }
        for (int i = 1; i < 10; i++)
        {
            tmp[tid, i] = (nmax[i - 1] - nmin[i - 1]) / (gmax[i - 1] - gmin[i - 1]) * (tmp[tid, i] - gmin[i - 1]) + nmin[i - 1];
        }
        test[tid] = tmp[tid, 1];

        tid = tid + thread.blockDim.x * thread.gridDim.x;
    }
}