쿠다 트리플 리만 합계

Question

나는 쿠다를 통해 트리플 리만 합계를 수행하려고합니다. 중첩 루프를 피하기 위해 합계 반복자에 다차원 격자 반복자를 사용하려고합니다. 중첩 된 커널을 사용할 수 없도록 2.0 테 살사 카드를 사용하고 있습니다.

필자가 필요로하는 각각의 x, y, z 변수에 대해 완전한 0 -> N 반복을 얻는 것으로 나타나지 않습니다. -----

int totalIterations = 128; // N value for single sum (i = 0; i < N) 
dim3 threadsPerBlock(8,8,8); 
dim3 blocksPerGrid((totalIterations + threadsPerBlock.x - 1)/threadsPerBlock.x, 
        (totalIterations + threadsPerBlock.y - 1)/threadsPerBlock.y, 
        (totalIterations + threadsPerBlock.z - 1)/threadsPerBlock.z); 
test<<<blocksPerGrid, threadsPerBlock>>>(); 

에 의해 호출

__global__ void test(){ 
uint xIteration = blockDim.x * blockIdx.x + threadIdx.x; 
uint yIteration = blockDim.y * blockIdx.y + threadIdx.y; 
uint zIteration = blockDim.z * blockIdx.z + threadIdx.z; 
printf("x: %d * %d + %d = %d\n y: %d * %d + %d = %d\n z: %d * %d + %d = %d\n", blockDim.x, blockIdx.x, threadIdx.x, xIteration, blockDim.y, blockIdx.y, threadIdx.y, yIteration, blockDim.z, blockIdx.z, threadIdx.z, zIteration); 
} 

---- ---- 출력 ----- 출력이 잘립니다

x y z 
... 
7 4 0 
7 4 1 
7 4 2 
7 4 3 
7 4 4 
7 4 5 
7 4 6 
7 4 7 
7 5 0 
7 5 1 
7 5 2 
7 5 3 
7 5 4 
7 5 5 
7 5 6 
7 5 7 
7 6 0 
7 6 1 
7 6 2 
7 6 3 
7 6 4 
7 6 5 
7 6 6 
7 6 7 
7 7 0 
7 7 1 
7 7 2 
7 7 3 
7 7 4 
7 7 5 
7 7 6 
7 7 7 
... 

, 지금 모든 무엇입니까 예를 들어,이 실행에서, 40 < z일 때 0 < x, y, z < 7 일 때x, y, z < 127이 필요합니다. 0 49, 0이어야합니다. < = z < = 127입니다. 다중 딤섬에 대한 이해는 잘못되었지만 리먼의 경우 반복기, x, y 및 z는 0에서 127 사이의 값을 가져야합니다.

또한 totalIterations> 128, 즉 1024를 만들면 프로그램이 cudaError 코드 6으로 종료됩니다.이 코드는 실행 타이머 만료로 이해합니다. 커널은 인쇄 작업 만하고 있기 때문에 시간 초과 이유를 모르겠습니다. 보조 장치에서이 기능을 실행하면 당장 문제가 해결 된 것으로 보입니다. 우리는 테슬라 중 하나를 사용하여 X를 실행하지만, geforce는 우편으로 새로운 디스플레이 장치가되어 계산 용 테슬라를 모두 비 웠습니다.

printf (...)는 합쳐질 함수의 실행으로 대체 될 것입니다.

생각은 나는 그것이 잠재적으로 거대한 (수백만 X 수백만의 X를 만들 메모리가 효율적으로 보이지 않는, 함수 값을 저장하는 방법을 잘 모르겠습니다 또한

for (int i = 0...) 
    for (int j = 0 ..) 
     for (int k = 0...) 

의 시리얼 코드 버전을 대체하는 것입니다 수백만) 3D 배열을 만들고이를 줄이거 나 함수의 값을 어떻게 든 공유 변수에 연결합니다.

는 ---- 장치 정보 (우리는 출력이 모두 동일합니다,이 카드 2X가 ---- 내가왔다으로 생각

Device 1: "Tesla C2050" 
    CUDA Driver Version/Runtime Version   5.0/5.0 
    CUDA Capability Major/Minor version number: 2.0 
    Total amount of global memory:     2687 MBytes (2817982464 bytes) 
    (14) Multiprocessors x (32) CUDA Cores/MP: 448 CUDA Cores 
    GPU Clock rate:        1147 MHz (1.15 GHz) 
    Memory Clock rate:        1500 Mhz 
    Memory Bus Width:        384-bit 
    L2 Cache Size:         786432 bytes 
    Max Texture Dimension Size (x,y,z)    1D=(65536), 2D=(65536,65535), 3D=(2048,2048,2048) 
    Max Layered Texture Size (dim) x layers  1D=(16384) x 2048, 2D=(16384,16384) x 2048 
    Total amount of constant memory:    65536 bytes 
    Total amount of shared memory per block:  49152 bytes 
    Total number of registers available per block: 32768 
    Warp size:          32 
    Maximum number of threads per multiprocessor: 1536 
    Maximum number of threads per block:   1024 
    Maximum sizes of each dimension of a block: 1024 x 1024 x 64 
    Maximum sizes of each dimension of a grid:  65535 x 65535 x 65535 
    Maximum memory pitch:       2147483647 bytes 
    Texture alignment:        512 bytes 
    Concurrent copy and execution:     Yes with 2 copy engine(s) 
    Run time limit on kernels:      No 
    Integrated GPU sharing Host Memory:   No 
    Support host page-locked memory mapping:  Yes 
    Concurrent kernel execution:     Yes 
    Alignment requirement for Surfaces:   Yes 
    Device has ECC support enabled:    Yes 
    Device is using TCC driver mode:    No 
    Device supports Unified Addressing (UVA):  Yes 
    Device PCI Bus ID/PCI location ID:   132/0 
    Compute Mode: 
    < Default (multiple host threads can use ::cudaSetDevice() with device simultaneously) > 

Answer 1

이미 장치 코드의 printf를 사용하는 것을 검증하는 것을 언급 x, y, z 배열의 모든 요소는 x, y, z 값이 큰 경우 현명하지 않습니다.

모든 요소 x, y , z는 스레드에 의해 접촉됩니다.

#include <stdio.h> 
#define DATAVAL 1 
#define cudaCheckErrors(msg) \ 
    do { \ 
     cudaError_t __err = cudaGetLastError(); \ 
     if (__err != cudaSuccess) { \ 
      fprintf(stderr, "Fatal error: %s (%s at %s:%d)\n", \ 
       msg, cudaGetErrorString(__err), \ 
       __FILE__, __LINE__); \ 
      fprintf(stderr, "*** FAILED - ABORTING\n"); \ 
      exit(1); \ 
     } \ 
    } while (0) 

__global__ void test(int *data, int dim){ 
    uint xIteration = blockDim.x * blockIdx.x + threadIdx.x; 
    uint yIteration = blockDim.y * blockIdx.y + threadIdx.y; 
    uint zIteration = blockDim.z * blockIdx.z + threadIdx.z; 

    data[((((zIteration*dim)+yIteration)*dim)+xIteration)]=DATAVAL; 
} 

int main(){ 
    int *testdata; 
    int *result; 
    int totalIterations = 128; // N value for single sum (i = 0; i < N) 
    int testsize = totalIterations*totalIterations*totalIterations; 
    dim3 threadsPerBlock(8,8,8); 
    dim3 blocksPerGrid((totalIterations + threadsPerBlock.x - 1)/threadsPerBlock.x, (totalIterations + threadsPerBlock.y - 1)/threadsPerBlock.y, (totalIterations + threadsPerBlock.z - 1)/threadsPerBlock.z); 
    cudaMalloc(&testdata, testsize*sizeof(int)); 
    cudaCheckErrors("cudaMalloc fail"); 
    cudaMemset(testdata, 0, testsize*sizeof(int)); 
    cudaCheckErrors("cudaMemset fail"); 
    result=(int *)malloc(testsize*sizeof(int)); 
    if (result == 0) {printf("malloc fail \n"); return 1;} 
    memset(result, 0, testsize*sizeof(int)); 
    test<<<blocksPerGrid, threadsPerBlock>>>(testdata, totalIterations); 
    cudaDeviceSynchronize(); 
    cudaCheckErrors("Kernel launch failure"); 
    cudaMemcpy(result, testdata, testsize*sizeof(int), cudaMemcpyDeviceToHost); 
    cudaCheckErrors("cudaMemcpy failure"); 

    for (unsigned i=0; i<testsize; i++) 
    if (result[i] != DATAVAL) {printf("fail! \n"); return 1;} 

    printf("Success \n"); 
    return 0; 

}