커널 템플릿과 함께 CUDA 커널 함수 포인터를받는 CUDA 런타임 API 함수를 사용하고 싶습니다.
템플릿없이 다음을 수행 할 수 있습니다.
__global__ myKernel()
{
...
}
void myFunc(const char* kernel_ptr)
{
...
// use API functions like
cudaFuncGetAttributes(&attrib, kernel_ptr);
...
}
int main()
{
myFunc(myKernel);
}
그러나 커널이 템플릿 인 경우 위의 내용은 작동하지 않습니다.
다른 예시:
#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <stdio.h>
template<typename T>
__global__ void addKernel(T *c, const T *a, const T *b)
{
int i = threadIdx.x;
c[i] = a[i] + b[i];
}
int main()
{
cudaFuncAttributes attrib;
cudaError_t err;
//OK:
err = cudaFuncGetAttributes(&attrib, addKernel<float>); // works fine
printf("result: %s, reg1: %d\n", cudaGetErrorString(err), attrib.numRegs);
//NOT OK:
//try to get function ptr to pass as an argument:
const char* ptr = addKernel<float>; // compile error
err = cudaFuncGetAttributes(&attrib, ptr);
printf("result: %s, reg2: %d\n", cudaGetErrorString(err), attrib.numRegs);
}
위의 결과 컴파일 오류가 발생합니다.
오류 : 필수 유형과 일치하는 함수 템플릿 "addKernel"의 인스턴스가 없습니다.
편집 : 지금까지 찾은 유일한 해결 방법은 myFunc (첫 번째 코드 예제 참조) 내부의 항목을 매크로에 넣는 것입니다. 이는보기 흉하지만 포인터 인수 전달이 필요하지 않으며 제대로 작동합니다.
#define MY_FUNC(kernel) \
{ \
...\
cudaFuncGetAttributes( &attrib, kernel ); \
...\
}
용법:
MY_FUNC( myKernel<float> )
"다른 예"에 포함 된 코드 참조 :
이것을 변경하십시오 :
const char* ptr = addKernel<float>; // compile error
이에:
void (*ptr)(float *, const float *, const float *) = addKernel<float>;
그리고 나는 그것이 올바르게 컴파일되고 실행될 것이라고 믿습니다.
나는 그것이 당신이하려는 일의 전반적인 범위에서 유용할지 여부를 모릅니다.
댓글의 질문에 응답하는 편집 :
함수에서 포인터를 "추출"하면 포인터를 다른 유형으로 캐스트 할 수 있습니다. 시도 해봐. 예를 들어 다음 코드도 작동합니다.
void (*ptr)(float *, const float *, const float *) = addKernel<float>;
const char *ptr1 = (char *)ptr;
err = cudaFuncGetAttributes(&attrib, ptr1);
따라서 질문에 답하기 위해 함수 포인터가 있으면 원하는 경우 함수 포인터를로 캐스팅 할 수 있습니다const char* .
그건 그렇고, 답변으로 게시 한 코드는 gcc 4.1.2 및 gcc 4.4.6에서 컴파일 오류를 발생시킵니다.
$ nvcc -arch=sm_20 -O3 -o t201 t201.cu
t201.cu: In function âint main()â:
t201.cu:25: error: address of overloaded function with no contextual type information
t201.cu:29: error: address of overloaded function with no contextual type information
$
그리고 &그 두 줄에서 제거하면 오류가 발생 합니다.
$ nvcc -arch=sm_20 -O3 -o t201 t201.cu
t201.cu: In function âint main()â:
t201.cu:25: error: insufficient contextual information to determine type
t201.cu:29: error: insufficient contextual information to determine type
$
따라서이 중 일부는 A 지점에서 B 지점으로 이동하는 데 필요한 단계가 컴파일러에 따라 달라질 수 있습니다.
이 기사는 인터넷에서 수집됩니다. 재 인쇄 할 때 출처를 알려주십시오.
침해가 발생한 경우 연락 주시기 바랍니다[email protected] 삭제
몇 마디 만하겠습니다