2012-09-22 2 views
2

CUDA에서 CTA가 할당 한 공유 메모리를 무료로 사용할 수있는 응용 프로그램 수준 API가 있습니까? CTA를 다른 작업에 다시 사용하고 이전 작업에서 사용한 메모리를 정리해야합니다.CUDA의 무료 공유 메모리

+0

공유 메모리가 커널 호출간에 지속되지 않습니다. 무엇이 문제입니까? 일반적으로 동일한 커널에서 더 많은 (이질적인) 작업을 수행하는 것보다 새로운 커널을 시작하는 것이 훨씬 낫습니다. –

+0

다른 작업에 동일한 CTA를 사용하여'__syncThreads()'와 동기화하여 다른 작업에 사용할 계획입니다. –

+0

Stefano, 커널 융합은 CUDA에서 커널 시작 오버 헤드를 상각하는 오래된 방법입니다. – ArchaeaSoftware

답변

4

공유 메모리는 커널 시작시 정적으로 할당됩니다. 선택적으로 커널에 크기를 지정하지 않은 공유 할당을 지정할 수 있습니다 많은 공유 메모리가 그 크기를 지정하지 않은 할당에 대응하는 방법을

__global__ void MyKernel() 
{ 
    __shared__ int fixedShared; 
    extern __shared__ int extraShared[]; 
    ... 
} 

세 번째 커널 실행 매개 변수는 다음을 지정합니다.

MyKernel<<<blocks, threads, numInts*sizeof(int)>>>(...); 

커널 발사 할당 공유 메모리의 총량은 커널 선언 양 플러스 공유 메모리 커넬 파라미터 더하여 정렬 헤드의 합이다. 커널을 실행하는 동안 할당 된 상태로 유지됩니다.

여러 단계의 실행을 거쳐 다른 목적으로 공유 메모리를 사용해야하는 커널의 경우 공유 메모리가있는 메모리를 다시 사용합니다. 크기가 지정되지 않은 선언에 포인터 연산을 사용합니다. 같은

뭔가 :

__global__ void MyKernel() 
{ 
    __shared__ int fixedShared; 
    extern __shared__ int extraShared[]; 
    ... 
    __syncthreads(); 
    char *nowINeedChars = (char *) extraShared; 
    ... 
} 

은 내가 threadFenceReduction 샘플이 __shared__ bool를 선언하지만,이 관용구를 사용하는 SDK 샘플을 알고도 감소의 부분 합계를 개최 공유 메모리를 사용하지 않습니다.

+1

커널 융합은 잘 작동하지만 레지스터 사용에주의해야합니다. 커널에 할당 된 레지스터의 수는 커널 시작과 동시에 정적이기 때문에 더 큰 커널의 레지스터 요구 사항이 높으면 커널 전체의 점유율이 줄어 듭니다. – ArchaeaSoftware