2016-10-05 1 views
1

병합되지 않은 메모리 액세스를 병합 된 것으로 변환하는 간단한 방법이 있는지 궁금합니다. 의이 배열의 예를 보자 :병합되지 않은 액세스에서 병합 된 메모리 액세스 CUDA

dW[[w0,w1,w2][w3,w4,w5][w6,w7][w8,w9]] 

지금, 나는 블록 0 액세스의 스레드 0 dW[0] 다음 블록 0 액세스 dw[1] 1 스레드 경우, 즉 글로벌 메모리에 합체 액세스 있다는 것을 알고있다. 문제는 두 가지 작업이 있다는 것입니다. 첫 번째 것은 위에서 설명한대로 병합됩니다. 그러나 두 번째 것은 블록 0의 스레드 1이 dW[0], dW[1]dW[2] 모두에서 연산을 수행해야하기 때문이 아닙니다.

컨테이너의 초기 모양이 병합 액세스를 허용하거나 금지한다는 것을 알고 있습니다. 그러나 dW은 매우 큰 배열이며 프로세스 중에 변환 할 수 없습니다.

이 문제를 해결할 수 있는지 알고 계십니까?

+1

(1) 벤치 마크 결과 벤치 마크 결과 미숙 한 메모리 액세스가 속도 저하의 중요한 원인이라고 알려 주셨습니까? (2) 아무도 메모리 액세스 패턴을 실제로 보여주는 [mcve]를 게시하지 않는 한 메모리 액세스를 최적화하는 데 도움을 줄 수 없습니다. –

+0

그래, 두 가지 벤치 마크를 수행했습니다. 그리고 그들은 경기 침체를 확인했습니다 (그리 많지는 않지만 천천히 ...). 이러한 작업은 복잡한 프로그램의 일부이기 때문에 항상 간단한 코드를 표시하는 것은 여전히 ​​어렵습니다. 나는 똑같은 문제를 일으키는 것을 구현하려고 노력할 것이다. –

답변

2

어쩌면 작동 할 수도있는 공유 메모리를 사용해 볼 수도 있습니다 (예를 들어 알지 못하더라도 어렵지 않게 말할 수 있습니다).

예를 들어, 첫 번째 작업은 병합 된 데이터에 액세스하고 두 번째 작업은 많이 걸음이라고 가정하십시오. 이 물건 속도를 빠르게 할 수 있습니다

아이디어는 회합 액세스가 공유 메모리에 문제가되지 않는다 (단, 뱅크 충돌은 다른 한편으로 할 수 있기 때문에, 수학을 위해 공유 사용 후 회합 방식으로 공유 데이터를로드하는 것입니다
__shared__ int shared[BLOCK_SIZE]; 
// Load data global -> shared with coalesced access ; you may need to load a bit more before/after depending on you application 
shared[tid] = global[some id] 
syncthreads(); 
// Do the math with coalescing access 
function0(shared[tid]) 
// Do the math with the non coalescing access 
function1(shared[tid+-1 or wathever]) 

손; 그것은 일반적으로 멋지다).

보다 정확한 도움말을 원하면 더 많은 정보를 제공해야합니다. 그건 힌트 일 뿐이야.

+0

감사합니다. 공유 메모리를 사용하려고합니다. 감사합니다. –

관련 문제