Ядро Cudafy не компилируется

Question

Я делаю первые шаги с Cudafy и пытаюсь написать функцию, которая будет принимать местоположение своего потока и на основе этого сохранять некоторое значение int в элемент массива. Мой код:

[Cudafy]
public static void GenerateRipples(GThread thread, int[] results)
{
  int threadPosInBlockX = thread.threadIdx.x;
  int threadPosInBlockY = thread.threadIdx.y;

  int blockPosInGridX = thread.blockIdx.x;
  int blockPosInGridY = thread.blockIdx.y;

  int gridSizeX = thread.gridDim.x;
  int gridSizeY = thread.gridDim.y;

  int blockSizeX = thread.blockDim.x;
  int blockSizeY = thread.blockDim.y;

  //int threadX = blockSizeX*blockPosInGridX + threadPosInBlockX;

  //if i use only one variable, everything is fine:
  int threadY = blockSizeY;

  //if i add or multiply anything, it cannot compile:
  //int threadY = blockSizeY*blockPosInGridY + threadPosInBlockY;


//  results[gridSizeX*blockSizeX*threadY + threadX] = 255;
}

Поэтому я не могу рассчитать threadY Вот. Если я использую более одной переменной в вычислениях, класс перевода Cudafy выдает ошибку (CudafyModule cm = CudafyTranslator.Cudafy(); создает исключение Cudafy.CudafyLanguageException).

Что я делаю неправильно?

ОБНОВЛЕНИЕ: Это код, который запускает ядро ​​на GPU:

public void RunTest2()
{
    GPGPU gpu = CudafyHost.GetDevice(CudafyModes.Target, CudafyModes.DeviceId);
    CudafyModule km = CudafyTranslator.Cudafy();
    gpu.LoadModule(km);

    int size = 20 * 20;
    int[] allPixels = new int[size];

    int[] dev_result = gpu.Allocate<int>(size);

    dim3 blocksInGrid = new dim3(5, 5);
    dim3 threadsPerBlock = new dim3(4, 4);

    gpu.Launch(blocksInGrid, threadsPerBlock).GenerateRipples(dev_result);
    gpu.CopyFromDevice(dev_result, allPixels);

    gpu.FreeAll();
}

Answer 1

Нам нужно посмотреть, как вы запускаете ваше ядро, приведенный выше код должен работать просто отлично. Я создал тестовый класс, который отлично работает и дает вам пример того, как подготовить размеры сетки / блока / потоков ядра. Если вы хотите увидеть отличные примеры, скачайте исходный код Cudafy и скомпилируйте проект CudafyExamples, посмотрите, как они готовят и используют функциональность CUDAfy.

** Примечание: я, должно быть, выкурил что-то очень хорошее, прежде чем отправлять первый класс, я забыл, чтобы убедиться, что он не вызывает нарушения доступа к памяти!!

Фиксированный класс ниже, нарушений нет.

Посмотрите отличные примеры на Codeproject и Stackru.

using System;
using System.Collections.Generic;
using System.Diagnostics;
using System.Linq;
using System.Text;

using Cudafy;
using Cudafy.Host;
using Cudafy.Translator;

namespace FxKernelTest 
{ 
    public class FxKernTest  
    {
        public GPGPU fxgpu;

        public const int N = 1024 * 64;

        public void ExeTestKernel()
        {
            GPGPU gpu = CudafyHost.GetDevice(CudafyModes.Target, 0);
            eArchitecture arch = gpu.GetArchitecture();
            CudafyModule km = CudafyTranslator.Cudafy(arch);

            gpu.LoadModule(km);

            int[] host_results = new int[N];

            // Either assign a new block of memory to hold results on device
            var dev_results = gpu.Allocate<int>(N);
            gpu.Set<int>(dev_results);

            // Or fill your array with values first and then
            for (int i = 0; i < N; i++) host_results[i] = i * 3;

            // Copy array with ints to device
            //var dev_filled_results = gpu.CopyToDevice(host_results);

            // 64*16 = 1024 threads per block (which is max for sm_30)
            dim3 threadsPerBlock = new dim3(64, 16);    

            // 8*8 = 64 blocks per grid, 1024 threads per block = kernel launched 65536 times
            dim3 blocksPerGrid = new dim3(8, 8); 

            //var threadsPerBlock = 1024; // this will only give you blockDim.x = 1024, .y = 0, .z = 0
            //var blocksPerGrid = 1;      // just for show

            gpu.Launch(blocksPerGrid, threadsPerBlock, "GenerateRipples", dev_results);

            gpu.CopyFromDevice(dev_results, host_results); 

            // Test our results
            for (int index = 0; index < N; index++)
                if (host_results[index] != index)
                    throw new Exception("Check your indexing math, genius!!!");
        }

        [Cudafy]
        public static void GenerateRipples(GThread thread, int[] results)
        {
            var blockSize = thread.blockDim.x * thread.blockDim.y;

            var offsetToGridY = blockSize * thread.gridDim.x;

            // This took me a few tries, I've never used 4 dimensions into a 1D array beofre :)

            var tid = thread.blockIdx.y * offsetToGridY +       // each Grid Y is 8192 in size
                      thread.blockIdx.x * blockSize +           // each Grid X is 1024 in size
                      thread.threadIdx.y * thread.blockDim.x +  // each Block Y is 64 in size
                      thread.threadIdx.x;                       // index into block


            var threadPosInBlockX = thread.threadIdx.x;

            var threadPosInBlockY = thread.threadIdx.y;

            var blockPosInGridX = thread.blockIdx.x;

            var blockPosInGridY = thread.blockIdx.y;

            var gridSizeX = thread.gridDim.x;

            var gridSizeY = thread.gridDim.y;

            var blockSizeX = thread.blockDim.x;

            var blockSizeY = thread.blockDim.y;

            // this is your code, see how I calculate the actual thread ID above!
            var threadX = blockSizeX * blockPosInGridX + threadPosInBlockX;

            //if i use only one variable, everything is fine:
            var threadY = blockSizeY;

            // this calculates just fine
            threadY = blockSizeY * blockPosInGridY + threadPosInBlockY;

            // hint: use NSight for Visual Studio and look at the NSight output, 
            // it reports access violations and tells you where...

            // if our threadId is within bounds of array size
            // we cause access violation if not
            // (class constants are automatically passed to kernels)
            if (tid < N)
                results[tid] = tid;

        }

    }
}

Информация ptxas: 0 байтов gmem Информация ptxas: Компиляция входной функции 'GenerateRipples' для 'sm_30' Информация ptxas: Свойства функции для GenerateRipples Строка стека 0 байт, хранилища разлива 0 байт, 0 разливов загружает информацию ptxas: Использовано 5 регистров, 328 байт cmem[0]