c++ - 理解 C++ 函数指针在 HIP 中通过引用传递

Question

HIP是对应NVIDIA的CUDA的AMD GPU编程模型。我有一个 code snippet来 self 无法完全理解的 HIP 源代码。提醒一下，理解以下代码片段不需要任何 HIP 背景知识，但更多的是 C++ 模板/函数指针中的问题。

typedef int hipLaunchParm;    
template <typename... Args, typename F = void (*)(hipLaunchParm, Args...)>
inline void hipLaunchKernel(F&& kernel, const dim3& numBlocks, const dim3& dimBlocks,
                        std::uint32_t groupMemBytes, hipStream_t stream, Args... args) 
{
    hipLaunchKernelGGL(kernel, numBlocks, dimBlocks, groupMemBytes, stream, 
        hipLaunchParm{}, std::move(args)...);
}

我对以下内容感到困惑:

• 如果 F 是一个函数指针，为什么它需要在参数中被双重引用？
• 第一个模板参数 typename...Args 有什么用？
• hipLaunchParm 只是整数的别名，但在参数中调用时 {} 是什么意思？

Answer 1

If F is a function pointer, why does it need to be double referenced in the argument?

F 不一定是函数指针。那只是默认类型。您可以传递任何 ^† 可调用对象，只要它可以使用给定参数调用，并且您希望避免在不必要时复制有状态函数对象。有些甚至可能无法复制。这可能就是他们在这里使用引用的原因。

^† 就 C++ 而言。我不知道 HIP/CUDA 可能有的限制。

How is the first template argument typename... Args useful?

它允许将可变数量的参数传递给委托(delegate)函数。

hipLaunchParm is just a alias for integer, but what is the meaning of {} when it is called in the argument?

T{} 是临时值初始化的语法。在整数的情况下，这意味着零作为参数传递。