2D 텍스처의 피치 정렬

Question

2D 텍스처는 이미지 프로세싱 애플리케이션에서 CUDA의 유용한 기능입니다. 피치 선형 메모리를 2D 텍스처에 바인딩하려면 메모리를 정렬해야합니다. cudaMallocPitch은 정렬 된 메모리 할당을위한 좋은 옵션입니다. 내 기기에서 cudaMallocPitch에 의해 반환 된 피치는 512의 배수입니다. 즉, 메모리는 512 바이트 정렬입니다.

장치의 실제 정렬 요구 사항은 내 장치에서 32 바이트 인 cudaDeviceProp::texturePitchAlignment에 의해 결정됩니다.

내 질문은 : 2D 텍스처의 실제 정렬 요구가 32 바이트 인 경우

는 왜 cudaMallocPitch 수익 512 바이트의 메모리를 정렬합니까?

메모리 낭비가 아닙니까? 예를 들어 크기가 513 x 100 인 8 비트 이미지를 만들면 1024 x 100 바이트를 차지합니다.

나는 다음과 같은 시스템에서이 동작을 얻을 :

1 : 아수스 G53JW + 윈도우 8 64 + 지포스 GTX 460M + CUDA 5 + 코어 i7 740QM + 4기가바이트 RAM

2 : 델 인스 파 이런 N5110 + 윈도우 7 64 + 지포스 GT525M + CUDA 4.2 + Corei7 2630QM + 6기가바이트 RAM

Answer 1

이 약간 투기 대답이지만, 할당의 피치 텍스처에 대한 만족해야 이 개 정렬 속성의 하나가 있다는 것을 염두에 두어야 textutr 포인터와 텍스처 행을위한 포인터. 나는 cudaMallocPitch이 cudaDeviceProp::textureAlignment에 의해 정의 된 전자를 기리는 것으로 의심합니다.

#include <cstdio> 

int main(void) 
{ 
    const int ncases = 12; 
    const size_t widths[ncases] = { 5, 10, 20, 50, 70, 90, 100, 
     200, 500, 700, 900, 1000 }; 
    const size_t height = 10; 

    float *vals[ncases]; 
    size_t pitches[ncases]; 

    struct cudaDeviceProp p; 
    cudaGetDeviceProperties(&p, 0); 
    fprintf(stdout, "Texture alignment = %zd bytes\n", 
      p.textureAlignment); 
    cudaSetDevice(0); 
    cudaFree(0); // establish context 

    for(int i=0; i<ncases; i++) { 
     cudaMallocPitch((void **)&vals[i], &pitches[i], 
      widths[i], height); 
     fprintf(stdout, "width = %zd <=> pitch = %zd \n", 
       widths[i], pitches[i]); 
    } 

    return 0; 
} 

는 A GT320M에서 다음을 제공합니다 : 예를 들어

Texture alignment = 256 bytes 
width = 5 <=> pitch = 256 
width = 10 <=> pitch = 256 
width = 20 <=> pitch = 256 
width = 50 <=> pitch = 256 
width = 70 <=> pitch = 256 
width = 90 <=> pitch = 256 
width = 100 <=> pitch = 256 
width = 200 <=> pitch = 256 
width = 500 <=> pitch = 512 
width = 700 <=> pitch = 768 
width = 900 <=> pitch = 1024 
width = 1000 <=> pitch = 1024 

나는 cudaDeviceProp::texturePitchAlignment가 배열에 적용 추측하고있다.

Answer 2

메모리 할당에 대한 실험을 마친 후 마침내 메모리를 절약 할 수있는 해결책을 찾았습니다. cudaMalloc으로 할당 된 메모리를 강제로 정렬하면 cudaBindTexture2D이 완벽하게 작동합니다.

cudaError_t alignedMalloc2D(void** ptr, int width, int height, int* pitch, int alignment = 32) 
{  
    if((width% alignment) != 0) 
     width+= (alignment - (width % alignment)); 

    (*pitch) = width; 

    return cudaMalloc(ptr,width* height); 
} 

이 함수에 의해 할당 된 메모리 cudaBindTexture2D의 요구 인 정렬 32 바이트이다. 내 메모리 사용량이 16 배 감소했으며 2D 텍스처를 사용하는 모든 CUDA 기능도 올바르게 작동합니다. 여기

현재 선택된 CUDA 장치 피치 정렬 요구 사항을 얻을 수있는 작은 유틸리티 기능입니다.

int getCurrentDeviceTexturePitchAlignment() 
{ 
    cudaDeviceProp prop; 
    int currentDevice = 0; 

    cudaGetDevice(&currentDevice); 

    cudaGetDeviceProperties(&prop,currentDevice); 

    return prop.texturePitchAlignment; 
}