调用函数指针 - 编程之家

如何解决调用函数指针

我正在编写一个可以在 GPU 上进行快速矩阵数学运算的库。它使用函数指针对矩阵中的数据进行操作。

我在 Visual Studio 2019 中使用 Cuda 11.2。

现在，我的问题是我用来对矩阵元素进行操作的内核崩溃了。

在调用为执行请求的操作而提供的函数指针时，似乎始终如一。当我在包装内核的函数中获取函数指针时，它不会崩溃。但是，如果我在内核所在的 .cu 文件之外的任何地方获取函数指针，它就会崩溃。它也只会在发布模式下崩溃，而不会在调试模式下崩溃。

我完全不知道这里发生了什么，有什么我忽略的地方吗？对我来说，这似乎是一个编译器错误，因为它只在发布模式下发生，而且只有当我在不同的 .cu 文件中获取函数指针时才会发生。

有什么办法可以解决我的问题吗？

这个崩溃也可以在不同的项目中重现，所以它不是一些旧的编译工件或其他东西。

我还做了一个最小的可重现示例：

内核.cu：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <stdio.h>
#include <math.h>
#include <iostream>

#include "crash.cuh"

template<class fnType,fnType fn>
static __global__ void __cudaFnPointerKernel(volatile void** fnPointer)
{
    auto fncopy = fn;
    *fnPointer = *((void**)&fncopy);
}

template<class fnType,fnType fn>
static inline decltype(fn) cudaFnPointer()
{
    volatile void** ptr = 0;
    cudamallocManaged(&ptr,sizeof(*ptr));

    __cudaFnPointerKernel<fnType,fn> << <1,1 >> > (ptr);
    volatile auto err = cudaDeviceSynchronize();

    decltype(fn) ptrcopy = (decltype(fn))*ptr;
    cudaFree(ptr);
    return ptrcopy;
}

__device__ float reLU(float x)
{
    return x > 0 ? x : 0;
}

__global__ void callFn(functionPointerType fn)
{
    fn(1);
}

int main()
{
    functionPointerConsumer consumer;

    auto fnPtr = cudaFnPointer<decltype(reLU)*,reLU>();
    std::cout << "fnPtr: " << fnPtr << "\n";

    callFn<<<1,1>>>(fnPtr);
    std::cout << "Kernel Error 1: " << cudaDeviceSynchronize() << "\n";

    consumer.crash(fnPtr);
    std::cout << "Kernel Error 2: " << cudaDeviceSynchronize() << "\n";
}

crash.cuh：

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <stdio.h>
#include <math.h>
#include <iostream>

typedef float(*functionPointerType)(float);

struct functionPointerConsumer
{
    void crash(functionPointerType fn);
};

crash.cu：

#include "crash.cuh"

__global__ void crashKernel(functionPointerType fn)
{
    fn(1);
    printf("Kernel did not crash\n");
}

void functionPointerConsumer::crash(functionPointerType fn)
{
    crashKernel<<<1,1>>>(fn);
}

调试模式下的预期输出/输出：

fnPtr: 00000007009A2E00
Kernel Error 1: 0
Kernel did not crash
Kernel Error 2: 0

释放模式下的实际输出/输出：

fnPtr: 0000000000000008
Kernel Error 1: 0
Kernel Error 2: 718

内核崩溃的错误代码是 718/cudaErrorInvalidPc。我不知道为什么它是 cudaErrorInvalidPc。