Почему это CUDA-ядро дает результаты, отличные от результатов CPU-версии?

Question

[Robotex]

Я портировал на CUDA этот кусок кода:

if(_layersCount > 1)
        {
            for(int i=_layersCount-2;i>=0;i--)
            {
                for(int j=0;j<_neuronsPerLayerCount[i];j++) // cuda kernel
                {
                    localGradients[indexByLayerAndNeuron(i, j)] = 0;
    
                    for(int k=0;k<_neuronsPerLayerCount[i+1];k++)
                    {
                        localGradients[indexByLayerAndNeuron(i, j)] += _neuronsInputsWeights[indexByLayerNeuronAndInput(i+1, k, j)]
                                                                        * localGradients[indexByLayerAndNeuron(i+1, k)];
                    }
    
                    localGradients[indexByLayerAndNeuron(i, j)] *= derivatives[indexByLayerAndNeuron(i, j)];
                }
            }
        }

Результат:

if(_layersCount > 1)
        {
            for(int i=_layersCount-2;i>=0;i--)
            {
                // calculateLocalGradientsForAnotherLayers
                blocksCount = floor((double) _neuronsPerLayerCount[i] / threads.x) + 1;
                blocks = dim3(blocksCount, 1);
    
                calculateLocalGradientsForAnotherLayers <<<blocks, threads>>> (deviceLocalGradients, _neuronsInputsWeights, deviceDerivatives, _neuronsPerLayerCount[i], _neuronsInPreviousLayers[i], _neuronsInPreviousLayers[i+1], _neuronsPerLayerCount[i+1], _inputsInPreviousLayers[i], _inputsInCurrentLayer[i]);
            }
        }

Ядро calculateLocalGradientsForAnotherLayers:

__global__ void calculateLocalGradientsForAnotherLayers(double * localGradients, double * neuronsInputsWeights, double * derivatives, int neuronsCount, int neuronsInPreviousLayers, int neuronsInPreviousLayersWithCurrent, int neuronsInNextLayer, int inputsInPreviousLayers, int inputsInCurrentLayer)
    {
        int idx = blockIdx.x * blockDim.x + threadIdx.x;
    
        if(idx < neuronsCount)
        {
            int neuron = neuronsInPreviousLayers + idx;
    
            localGradients[neuron] = 0;
    
            // this to Kernel, then reduce localGradients.
            for(int k=0;k<neuronsInNextLayer;k++)
            {
                localGradients[neuron] += neuronsInputsWeights[inputsInPreviousLayers + k*inputsInCurrentLayer + idx]
                                                                * localGradients[neuronsInPreviousLayersWithCurrent + k];
            }
    
            localGradients[neuron] *= derivatives[neuron];
        }
    }

Но я вижу различия в результатах вычислений, начиная со второго знака после запятой. В чем может быть проблема? Все остальные ядра за исключением этого работают отлично.


Моя карта GF555M. Она поддерживает двойную точность.

Answer 1

[oleksandr_veles]

Возможно проблема в том, что ядра в цикле выполняются асинхронно.
А результат выполнения одного ядра зависит от результата другого, и следующее ядро в цикле выполняется не дождавшись окончания выполнения предыдущего.
Двойную точность, если у Вас не тесла к20 или старые тесла 2ххх, смысла использовать нет на картах nvidia.
GF555M и в одинарной точности не блещет особо, а с двойной ещё на порядок медленнее.
Cовременный 4 ядерный CPU будет быстрее в double.

Answer 2

[Robotex]

Я переписал ядро так, чтобы оно симулировалось циклами на CPU, чтобы проверить проблема ли в GPU. Погрешность осталась, следовательно ошибка в коде скорее всего. Ищу ее…

Comments

[Robotex]

Так и есть. Оказалось, банальная опечатка. В строке:

calculateLocalGradientsForAnotherLayers <<<blocks, threads>>> (deviceLocalGradients, _neuronsInputsWeights, deviceDerivatives, _neuronsPerLayerCount[i], _neuronsInPreviousLayers[i], _neuronsInPreviousLayers[i+1], _neuronsPerLayerCount[i+1], _inputsInPreviousLayers[i], _inputsInCurrentLayer[i]);

следовало написать:

calculateLocalGradientsForAnotherLayers <<<blocks, threads>>> (deviceLocalGradients, _neuronsInputsWeights, deviceDerivatives, _neuronsPerLayerCount[i], _neuronsInPreviousLayers[i], _neuronsInPreviousLayers[i+1], _neuronsPerLayerCount[i+1], _inputsInPreviousLayers[i+1], _inputsInCurrentLayer[i+1]);