模板函数调用其他函数

Question

我了解template functions 通常要在头文件中声明和定义。

我遇到的问题是我的template function 调用了其他函数。这些其他函数的原型位于模板函数本身之前的同一个头文件中。

那部分代码：

//header.h

template <int ignoreAdetection>
__global__ void MCMLKernel(SimState d_state, GPUThreadStates tstates)
{
  // photon structure stored in registers
  PhotonStructGPU photon;

  // random number seeds
  UINT64 rnd_x;
  UINT32 rnd_a;

  // Flag to indicate if this thread is active
  UINT32 is_active;

  // Restore the thread state from global memory.
  RestoreThreadState(&d_state, &tstates, &photon, &rnd_x, &rnd_a, &is_active);
...
...
}

函数RestoreThreadState 是从这个模板函数调用的几个函数中的第一个。其他的在 for 循环中调用。

我不确定这个模板函数是否应该在头文件中。如果应该在头文件中，我该如何调用其他函数？

我在 MCMLKernel 实例化期间从编译器得到的错误：

• 错误：缺少显式类型（假定为“int”）
• 错误：变量“RestoreThreadState”可能未初始化
• 错误：“SimState *”类型的值不能用于初始化“int”类型的实体
• 错误：应为“）”
• 警告：声明与之前的“RestoreThreadState”不兼容

其他细节。所有这些函数都是 CUDA 内核函数。 MCMLKernel 是一个 __global__ 内核，它调用的其余函数是 __device__ 内核。我正在使用 Nsight Eclipse Edition 和计算能力 1.3 GPU（四个 Tesla C1060 卡）。

Answer 1

__device__ void RestoreThreadState(...); 是前向声明，意味着在同一个翻译单元中遵循正确的定义（它甚至可能在函数调用之下）。

__device__ void RestoreThreadState(...) {} 是一个无操作函数的定义。

__device__ int HitBoundary(PhotonStructGPU *photon); 是一个前向声明。你提供定义吗？

更新

Here's the Gist 基于您提供的代码 sn-ps。我将内核、设备函数和类型声明保存在 3 个单独的标头中。我可以编译它（虽然显然它不能运行）

Answer 2

@Eugene 提供了这个问题的答案。

我这样创建了被调用函数的原型

__device__ void RestoreThreadState(SimState *d_state, GPUThreadStates *tstates,
                               PhotonStructGPU *photon,
                               UINT64 *rnd_x, UINT32 *rnd_a,
                               UINT32 *is_active);

但是，用尤金给出的一个例子，它应该是这样的（不是原型！见其他答案）

__device__ void RestoreThreadState(SimState *d_state, GPUThreadStates *tstates,
                               PhotonStructGPU *photon,
                               UINT64 *rnd_x, UINT32 *rnd_a,
                               UINT32 *is_active){}

注意末尾的{}。

接下来在代码中我调用了实际上返回类似这样的值的函数

__device__ int HitBoundary(PhotonStructGPU *photon);

从同一个 MCMLKernel 模板函数调用。它给了我一个警告：

• 警告：在非空函数“HitBoundary”的末尾缺少返回语句

更新：在另一个源文件 kernel.cu 我有以下声明和定义（我有多个定义问题的原因）：

__device__ void RestoreThreadState(SimState *d_state, GPUThreadStates *tstates,
                                   PhotonStructGPU *photon,
                                   UINT64 *rnd_x, UINT32 *rnd_a,
                                   UINT32 *is_active)
{
  UINT32 tid = blockIdx.x * NUM_THREADS_PER_BLOCK + threadIdx.x;

  *rnd_x = d_state->x[tid];
  *rnd_a = d_state->a[tid];

  photon->x = tstates->photon_x[tid];
  photon->y = tstates->photon_y[tid];
  photon->z = tstates->photon_z[tid];
  photon->ux = tstates->photon_ux[tid];
  photon->uy = tstates->photon_uy[tid];
  photon->uz = tstates->photon_uz[tid];
  photon->w = tstates->photon_w[tid];
  photon->sleft = tstates->photon_sleft[tid];
  photon->layer = tstates->photon_layer[tid];

  *is_active = tstates->is_active[tid];
}

总结：我有四个源文件

1. ma​​in.cu
2. kernel.cu
3. rng.cu
4. mem.cu

每个源文件，除了 ma​​in.cu 都有一个关联的标题

1. kernel.cuh
2. rng.cuh
3. mem.cuh

我想在 ma​​in.cu 中前向声明要使用的函数。

一切都很好，直到我到达调用 kernel.cu 和 rng.cu 的函数的 模板函数。