cuModuleGetFunction返回未找到

我想用nvrtc JIT編譯器來編譯CUDA內核來提高我的應用程序的性能（所以我有更多的指令提取但我保存了多個數組訪問）。cuModuleGetFunction返回未找到

這些函數看起來像例如這樣，由我的函數發生器（不那麼重要）產生：

extern "C" __device__ void GetSumOfBranches(double* branches, double* outSum) 
{ 
    double sum = (branches[38])+(-branches[334])+(-branches[398])+(-branches[411]); 
    *outSum = sum; 
}

我編譯上面用下面的函數代碼：

CUfunction* FunctionGenerator::CreateFunction(const char* programText) 
{ 
     // When I comment this statement out the output of the PTX file is changing 
     // what is the reson?! 
     // Bug? 
     std::string savedString = std::string(programText); 


     nvrtcProgram prog; 
     nvrtcCreateProgram(&prog, programText, "GetSumOfBranches.cu", 0, NULL, NULL); 

     const char *opts[] = {"--gpu-architecture=compute_52", "--fmad=false"}; 
     nvrtcCompileProgram(prog, 2, opts); 

     // Obtain compilation log from the program. 
     size_t logSize; 
     nvrtcGetProgramLogSize(prog, &logSize); 
     char *log = new char[logSize]; 
     nvrtcGetProgramLog(prog, log); 
     // Obtain PTX from the program. 
     size_t ptxSize; 
     nvrtcGetPTXSize(prog, &ptxSize); 
     char *ptx = new char[ptxSize]; 
     nvrtcGetPTX(prog, ptx); 

     printf("%s", ptx); 

     CUdevice cuDevice; 
     CUcontext context; 
     CUmodule module; 
     CUfunction* kernel; 
     kernel = (CUfunction*)malloc(sizeof(CUfunction)); 
     cuInit(0); 
     cuDeviceGet(&cuDevice, 0); 
     cuCtxCreate(&context, 0, cuDevice); 
     auto resultLoad = cuModuleLoadDataEx(&module, ptx, 0, 0, 0); 
     auto resultGetF = cuModuleGetFunction(kernel, module, "GetSumOfBranches"); 
     return kernel; 
}

一切工作，除了cuModuleGetFunction正在恢復CUDA_ERROR_NOT_FOUND。出現此錯誤的原因是在PTX文件中找不到GetSumOfBranches。

然而printf("%s", ptx);輸出是這樣的：

// Generated by NVIDIA NVVM Compiler 
// 
// Compiler Build ID: CL-19856038 
// Cuda compilation tools, release 7.5, V7.5.17 
// Based on LLVM 3.4svn 
// 

.version 4.3 
.target sm_52 
.address_size 64 

    // .globl GetSumOfBranches 

.visible .func GetSumOfBranches(
    .param .b64 GetSumOfBranches_param_0, 
    .param .b64 GetSumOfBranches_param_1 
) 
{ 
    .reg .f64 %fd<8>; 
    .reg .b64 %rd<3>; 


    ld.param.u64 %rd1, [GetSumOfBranches_param_0]; 
    ld.param.u64 %rd2, [GetSumOfBranches_param_1]; 
    ld.f64 %fd1, [%rd1+304]; 
    ld.f64 %fd2, [%rd1+2672]; 
    sub.rn.f64 %fd3, %fd1, %fd2; 
    ld.f64 %fd4, [%rd1+3184]; 
    sub.rn.f64 %fd5, %fd3, %fd4; 
    ld.f64 %fd6, [%rd1+3288]; 
    sub.rn.f64 %fd7, %fd5, %fd6; 
    st.f64 [%rd2], %fd7; 
    ret; 
}

在我optinion一切都很好，並GetSumOfBranches通過前人的精力來cuModuleGetFunction發現。你能解釋我爲什麼嗎？

第二個問題

當我outcomment std::string savedString = std::string(programText);那麼PTX的輸出就是：

// Generated by NVIDIA NVVM Compiler 
// 
// Compiler Build ID: CL-19856038 
// Cuda compilation tools, release 7.5, V7.5.17 
// Based on LLVM 3.4svn 
// 

.version 4.3 
.target sm_52 
.address_size 64

因爲savedString完全不使用這是奇怪...

來源

2016-09-27 Jens

爲什麼你有標記爲'__device__'的問題？這不是CUDA內核。你不能從主機代碼中調用它。無論如何，對於類型問題「爲什麼這不起作用？」因此你應該提供一個[mcve] –

我沒有想過我只編譯內核..我認爲我也可以編譯設備函數。你剛剛解決了我的問題（也許你想回答我的問題以獲得積分？ - 否則謝謝你）。在下一個問題中，我將提供一個最小的，完整的和可驗證的例子。 – Jens

你所要做的不支持。主機端模塊管理API和設備ELF格式不公開__device__函數，只有__global__函數可通過內核啓動API調用。

您可以事先或在運行時編譯設備函數，並以JIT方式將它們與內核鏈接起來，然後您可以檢索這些內核並調用它們。但這就是你所能做的。

來源

2016-09-27 17:02:42 talonmies

cuModuleGetFunction返回未找到

回答

相關問題