MATLAB로 CUDA 프로그램을 만드는 방법에 대한 제안입니다.
이처럼 전례 없는 오리지널 처리는 cuda에서 조합된 경우 변수별 동작과 배열 외 참조 등을 사전에 확인하는 경우가 많다.저는 그 자체가 모든 언어가 똑같다고 생각합니다. 특히cuda는 병행 처리가 기본이고 디버깅 도구가 좋지 않아서 어디가 틀렸는지 발견하기 어려워서 최근까지 어려움을 겪었습니다.
이advent 달력을 주최하는@motorcontrolman씨는Simultink 과격파로 활동하고 있습니다. 하지만 저는 MATLAB 원리주의자이기 때문에 MATLAB로cuda의 알고리즘 계획을 세우는 방법을 고려했습니다.나는 누구에게 수요가 있는지 모르겠지만, 나는 절반은 나를 위해 여기에 쓸 것이다.
1. MATLAB Coder 사용 여부
빠른 cuda를 원한다면 MATLAB Coder를 사용할 수 있습니다.느린 gpu로 이루어지면 되지만 속도를 추구하는 사람은 추천하지 않습니다.gpu 인코딩과 matlab 같은 cpu 인코딩은 사상이 완전히 다른 부분이 있고 메모리 복제 등 최적화되기 전까지는 하지 않기 때문이다.
사상이 다르면 예를 들어 매틀라비는 자신이 다음과 같은 화식 처리를 하고 싶어 한다.
입력:a=[a1,a2,a3,...,aN]
출력: b=a1+...+aN
cpu 기반 인코딩이라면 이렇게 하겠죠.
% a, Nは定義済みとする。
for idx = 1:N
b = b + a(idx)
end
※ 마트라비라면 썸()으로 쓸 수 있습니다!아주 편리하군요!
하지만 GPU는 이것마저도 간단해 보이는 처리에 공을 들여야 한다.상술한 방법에서 한 라인이 수조의 참조 위치를 이동하면서 순서대로 덧셈 연산을 하기 때문에 계산량은 $O(N)$이다.이렇게 되면 전혀 병행화할 수 없기 때문에 GPU는 다른 방법을 취할 것이다.
16개 요소의 배열을 합칠 때 그림에서 보듯이 먼저 이웃한 동지를 더한 다음에 두 개의 옆집을 더한다.이렇게 하면 $O(\log2(N)달러의 주문 계산으로 끝납니다.
매틀랩 코더는 이 틈을 채우지 않는 일이 많고 CPU-GPU 간 메모리 전송 등을 최적화하지 않아 결국 서서히 이뤄지고 있기 때문이다.
2. 어쨌든 큐다의 내용을 매틀라비로 기술하고 싶다!
MATLAB가 아니더라도 단식홍의 프로그래밍은 이 사상의 차이를 자동으로 메우지 못하기 때문에 나는 이런 업무 흐름으로 CUDA를 조합한다.
2-1. Block 순환과thread 순환을 하는 이중 순환 시스템
GPU의 병렬 인덱스는 블록과 스레드의 개념이 있어 좀 복잡하다.스레드는 여러분이 상상하는 것과 같이 처리하는 단위이고 블록은 스레드(최근 1024스레드)를 묶는 것입니다.모든 블록과 루틴에는 색인이 있지만, 루틴 색인은 일반적인 그림이 아니라 블록에 폐쇄된 색인입니다.
따라서 일반적으로 CUDA 커널에 들어간 후 각 처리에 대응하는 일대일 스레드 인덱스를 계산합니다.
__global__ void Hoge(int num, T* a){
// blockDim.xはブロック内のスレッド総数
const int idx = blockDim.x * blockIdx.x + threadIdx.x;
if(idx >= num){
return;
}
// 以下何かの計算
}
코드를 보면 알 수 있듯이 블록과 라인은 모두 3차원 값이 있기 때문에 그때는 번거로운 계산이 필요하다.그리고 블록 인덱스와 라인 인덱스를 독립적으로 사용하는 경우도 있기 때문에 단순히 대량으로 병행하는 생각과는 다른 생각을 해야 한다.
MATLAB를 사용하여 GPU의 특수성을 시뮬레이션하는 방법은 다음과 같습니다.
a = [1, 2, 3, 4, 5];
b = [6, 7, 8, 9, 10];
a = a + b;
b = a + b;
GPU에 동기를 의식하지 않고 쓰면 a=a+b까지 반복해서 읽지 않기 때문에 순조롭게 진행되지만 어떤 라인에서 a=a+b를 진행할 때 다른 라인은 b=a+b를 진행하기 때문에 GPU에서 이 시간 안에 명확하게 동기화해야 한다.2. CUDA에서는 BlockIdx와threadIdx로 배열된 인덱스 idx = BlockDim*BlockIdx+threadIdx를 구해야 하지만 앞에서 말한 바와 같이 BlockIdx와threadIdx를 단독으로 사용할 때도 있기 때문에 쉽게 사용할 수 있기를 바랍니다.
3. 이 점에 대해 MATLAB가 하고 싶은 일은 GPU 최적화가 아니라 알고리즘 검증일 뿐이다.CUDA에서 방금 전의 동기화는 몇 가지 유형(block 내에서만 동기화되는지 GPU 전체에서 동기화되는지 등)이 있지만 어떤 동기화를 사용하는 것이 최적화된 분야이기 때문에 MATLAB는 구별할 필요가 없다.
이를 성립시키는 좋은 방법은 Block 순환과thread 순환의 이중 순환 시스템을 만드는 것이다.코드상으로는 이런 느낌입니다.
% 扱う配列の要素数
N = 1000;
blockDim = 512;
blockNum = floor((N + blockDim - 1) / blockDim);
for blockIdx = 0:blockNum - 1
for threadIdx = 0:blockDim - 1
% forの中身がほぼそのままCUDAカーネルになる。
% CUDAでよく使うインデックスの算出方法をそのまま使える。
idx = blockDim * blockIdx + threadIdx;
% 以下なんかしらの処理
end
end
이 방법의 장점은 쌍for 내에서 CUDA 커널에 가깝다는 기술이다.만약 이렇게 검증이 가능하다면 for 안의 이미지를 기본적으로 원형대로 CUDA에 촬영하면 작동할 수 있다.근데 색인 같은 거 주의해야 돼.처음 표시된 요점을 충족하는지 확인
루틴 사이의 동기화를 보장하지 않습니다. (그룹 요소와 다시 읽을 수 있습니다.)
for 순환이기 때문에 진열 요소 간의 동기화는 보장되지 않을 뿐만 아니라 완전히 동기화되지 않습니다.깨끗이 제거하다.
BlockIdx와threadIdx는 특수 처리 없이 사용할 수 있습니다.
BlockIdx의 배열이나 threadIdx의 배열을 만들지 않았기 때문에 잘 알고 있습니다.
나는 알고리즘을 철저히 검증하고 싶다.
늘 걱정했던 동기에 대해서는 for로 나눠서 할 수 있어 나쁠 게 없었다.
3. 문제가 좀 있습니다.
for 루프를 이중으로 설정하면 MATLAB에서도 CUDA를 검증할 수 있습니다.
하지만 다 좋은 건 아니야.여기서 주의해야 할 점과 좋지 않은 점.
3-1. 여보세요.
CUDA를 CUPU로만 조합해보면 시키는 일이 너무 느리고, MATLAB 같은 해석기형 언어로 조합하면 당연히 느리다.그리고 MATLAB가 잘하는 배열 계산은 일부러 for로 나뉘기 때문에 더욱 느리다...이곳은 좋지 않은 곳이다.
3-2. 인덱스 시작 1
CUDA는 c++를 따르기 때문에 인덱스는 0으로 시작합니다.MATLAB로 검증하면 가끔 혼란스러워질 수 있어요.되도록 0부터 기술하고 읽기와 쓰기 배열할 때만 +1으로 회피한다.
4. 장점
여기까지 들으니 c++나pythhon의 디버깅 실행과 다를 것 같지만, MATLAB의 난이도는 매우 낮습니다.좋은 곳도 소개해 주세요.
4-1. 그래프 따위는 쉽게 꺼낼 수 있다.
MATLAB의 풍부한 그래픽 클러스터를 사용하면 GPU 내부의 움직임을 시각화할 수 있습니다.중복 읽기와 쓰기가 있는지 바로 확인할 수 있어 편리하다.
4-2.문서화할 수 있습니다.
현장 편집기에서 조립하면 공식과 도형의 해설서를 마음대로 만들 수 있다.알고리즘 연구-테스트 실현-검증-해설 문서 작성 절차 모두 완료.노력이 격감하다.
4-3. 최후
그럼에도 불구하고 색인 1의 시작은 뒷걸음질쳤다.좋은 방법 있으면 알려주세요.
Reference
이 문제에 관하여(MATLAB로 CUDA 프로그램을 만드는 방법에 대한 제안입니다.), 우리는 이곳에서 더 많은 자료를 발견하고 링크를 클릭하여 보았다 https://qiita.com/tommyecguitar/items/e268b012adea814beda1텍스트를 자유롭게 공유하거나 복사할 수 있습니다.하지만 이 문서의 URL은 참조 URL로 남겨 두십시오.
우수한 개발자 콘텐츠 발견에 전념 (Collection and Share based on the CC Protocol.)