2012-03-30 3 views
2

다음과 같은 계산이 필요합니다. A [x] [y] = sum {z = 0에서 z = n까지} {B [x] [y] [z] + C [x] y] [z]}의 행렬 A는 차원 [높이] [너비]와 행렬 B를 가지며 C는 차원 [높이] [너비] [n]을가집니다.합계 3D 행렬

index = 0; 
for (z = 0; z<n; ++z) 
    for(y = 0; y<width; ++y) 
     for(x = 0; x<height; ++x) { 
      matrix[index] = value; 
      index++; 
     } 

Q1 :

값이 같은 뭔가 메모리에 매핑 된이 쿠다 커널은 괜찮아?

idx = blockIdx.x*blockDim.x + threadIdx.x; 
idy = blockIdx.y*blockDim.y + threadIdx.y; 

for(z=0; z<n; z++){ 
    A[idx*width+idy] += B[idx*width+idy+z*width*height] + C[idx*width+idy+z*width*height]; 
} 

질문 2 : 계산을 더 빨리 할 수 ​​있습니까?

idx = blockIdx.x*blockDim.x + threadIdx.x; 
idy = blockIdx.y*blockDim.y + threadIdx.y; 
idz = blockIdx.z*blockDim.z + threadIdx.z; 

int stride_x = blockDim.x * gridDim.x; 
int stride_y = blockDim.y * gridDim.y; 
int stride_z = blockDim.z * gridDim.z; 

while (idx < height && idy < width && idz < n) { 
    atomicAdd(&(A[idx*width+idy]), B[idx*width+idy+idz*width*height] + C[idx*width+idy+idz*width*height]); 
    idx += stride_x; 
    idy += stride_y; 
    idz += stride_z; 
} 

답변

1

Q1 : 시험 그것은 당신이 대답을

비를 알고 행렬과 : 당신은 문제가 발생할 수 있습니다 매우 큰 행렬을 사용하는 경우. 적절한 증분으로 while 루프를 사용하십시오. Example by Cuda는 평소처럼 참고서입니다.

중첩 루프를 구현하는 예제는 For nested loops with CUDA입니다. 거기에 while 루프가 구현됩니다.

marina.k는 경쟁 조건에 대한 권리입니다. 그것은 원자 적 연산이 코드를 느리게하는 경향이 있기 때문에 접근법 1을 선호합니다.

+0

최대 차수 = 2 (CC 대 <2) = 3 (CC 용> = 2). 그래서 나는 문제가 없다고 생각한다. 나는 큰 행렬을 가지고 있지만 문제를 보지 못한다. – user1281071

+0

알았어. 나는 나의 대답을 업데이트했다. – Azrael3000

+0

당신은 다음을 명심하십시오.'if (idx> = height || idy> = width || idz> = n) return;'? – user1281071

2

첫 번째 커널은 정상입니다. 그러나 행렬 BC에 대한 액세스를 병합하지 않았습니다.

두 번째 커널 기능과 동일합니다. 당신은 하나의 스레드가 A[idx*width+idy] 주소에 쓸 수있는 능력을 가지고 있기 때문에 데이터 레이싱의 원인이 있습니다. AttomicAdd

일반 질문 : 나는 실험 결과 더 좋다고 생각합니다. 그것은 당신이 가지고있는 전형적인 매트릭스 크기에 달려 있습니다. Fermi의 최대 스레드 블록 크기는 < 1024이고 행렬의 크기가 클 경우 많은 스레드 블록을 사용할 수 있다는 점을 기억하십시오. 일반적으로 스레드 블록이 많은 경우 속도가 느립니다. ArrayFire

+0

데이터 레이싱, thx로 좋은 점 :) 원자 함수를 추가 할 때 코드 양식 Q2는 Q1에서 코드와 정확히 동일하게 수행합니까? 내 주요 목표는 속도이므로 나는 가장 빠른 해결책을 찾고 있습니다. – user1281071

+0

내 대답을 편집했습니다. – geek

+0

"실험은 더 낫다는 것을 보여줍니다", 원자 함수를 사용하는 것이 더 낫습니다. 사용하지 않습니까? – user1281071

2

실시간 단순 : 나사 블록 = 3 스레드 블록 격자의 최대 차수의

array A = randu(nx,ny,nz); 
array B = sum(A,2); // sum along 3rd dimension 
print(B);