PyCUDA: синтаксис для функции, которая вызывает функцию

Question

При использовании функции из SourceModule, которая зависит от другой функции в SourceModule, как я передаю ее в вызове функции, т.е. что такое "???" в следующем коде:

import numpy
import pycuda.autoinit
import pycuda.driver as drv
from pycuda.compiler import SourceModule

mod = SourceModule("""
__global__ void make_square(float *in_array, float *out_array)
{
  int i;
  int N = 5;
  for (i=0; i<N; i++)
  {
    out_array[i] = pow(in_array[i],2);
  }
}
__global__ void make_square_add_one(float *in_array, float *out_array, void make_square(float *, float *))
{
  int N = 5;
  make_square(in_array,out_array);
  for (int i=0; i<N; i++)
    out_array[i] = out_array[i] + 1;
}
""")

make_square = mod.get_function("make_square")
make_square_add_one = mod.get_function("make_square_add_one")
in_array = numpy.array([1.,2.,3.,4.,5.]).astype(numpy.float32)
out_array = numpy.zeros_like(in_array).astype(numpy.float32)
make_square_add_one(drv.In(in_array), drv.Out(out_array), ??? , block = (1,1,1), grid = (1,1))

Спасибо за любую информацию.

Answer 1

В традиционной модели выполнения CUDA функции __global__ являются ядрами, и они не могут передаваться в качестве аргументов другим ядрам и не могут вызываться другими ядрами. Это выглядит как make_square действительно должно быть функцией устройства, что-то вроде:

__device__ void make_square(float *in_array, float *out_array)
{
  int i;
  for (i=0; i<5; i++)
  {
    out_array[i] = pow(in_array[i],2);
  }
}

который затем вызывается из запущенного ядра как:

__global__ void make_square_add_one(float *in_array, float *out_array)
{
  int N = 5;
  make_square(in_array,out_array);
  for (int i=0; i<N; i++)
    out_array[i] = out_array[i] + 1;
}

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