묻고 답해요
156만명의 커뮤니티!! 함께 토론해봐요.
인프런 TOP Writers
-
미해결CUDA 프로그래밍 (2) - C/C++/GPU 병렬 컴퓨팅 - 벡터 vector 더하기
memcpy 시간 고려 시 gpu vs. cpu
안녕하세요.올려주신 강의 잘 듣고 있습니다.memcpy 시간까지 고려하면 cpu로 연산한 것이 더 빠른 것 같은데 맞나요?(1분 51초)감사합니다.
-
미해결CUDA 프로그래밍 (1) - C/C++/GPU 병렬 컴퓨팅 - CUDA 커널 kernel
cudaFree가 자동으로 되는 상황
안녕하세요. 좋은 강의 올려주셔서 정말 감사합니다. 39강 8분 14초에서 cudaFree를 하지 않아도 프레임이 다 종료될때 자동으로 free가 된다고 하셨는데, 프레임이 종료된다는 것이 어떤 의미인지 잘 모르겠습니다. 감사합니다.
-
미해결CUDA 프로그래밍 (1) - C/C++/GPU 병렬 컴퓨팅 - CUDA 커널 kernel
MSB3721 오류
안녕하세요 쿠다 실행하던중에 삭제하고 다운해도 계속 같은 오류가 반복돼서 해결방법을 아실까 하고 질문 드립니다. 심각도 코드 설명 프로젝트 파일 줄 비표시 오류(Suppression) 상태오류 MSB3721 ""C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.9\bin\nvcc.exe" -gencode=arch=compute_52,code=\"sm_52,compute_52\" --use-local-env -ccbin "C:\Program Files\Microsoft Visual Studio\2022\Community\VC\Tools\MSVC\14.34.31933\bin\HostX64\x64" -x cu -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.9\include" -I"C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v12.9\include" --keep-dir x64\Release -maxrregcount=0 --machine 64 --compile -cudart static -DWIN32 -DWIN64 -DNDEBUG -D_CONSOLE -D_MBCS -Xcompiler "/EHsc /W3 /nologo /O2 /FS /MD " -Xcompiler "/Fdx64\Release\vc143.pdb" -o D:\M&S\LAB\STUDY\cuda\cuda2\CudaRuntime1\CudaRuntime1\x64\Release\kernel.cu.obj "D:\M&S\LAB\STUDY\cuda\cuda2\CudaRuntime1\CudaRuntime1\kernel.cu"" 명령이 종료되었습니다(코드: 1). CudaRuntime1 C:\Program Files\Microsoft Visual Studio\2022\Community\MSBuild\Microsoft\VC\v170\BuildCustomizations\CUDA 12.9.targets 801 환경변수도 이렇게 다 했는데 안됩니다.감사합니다.
-
해결됨CUDA 프로그래밍 (1) - C/C++/GPU 병렬 컴퓨팅 - CUDA 커널 kernel
식별자 threadIdx 정의 되지 않음 오류
- 출장이 잦아서, 답변에 시간이 좀 걸릴 수도 있습니다. 양해 바랍니다. - 학습 관련 질문을 남겨주세요. 상세히 작성하면 더 좋아요! - 먼저 유사한 질문이 있었는지 검색해보세요. - 서로 예의를 지키며 존중하는 문화를 만들어가요. - 잠깐! 인프런 서비스 운영 관련 문의는 1:1 문의하기를 이용해주세요. 선생님 강의를 최대한 따라 해보았는데 실행은 잘되고 cuda success까지 잘 나오는데 오류랑 경고가 없어지지 않아 질문 남기게 되었습니다.
-
미해결CUDA 프로그래밍 (1) - C/C++/GPU 병렬 컴퓨팅 - CUDA 커널 kernel
cudaGetLastError() 에러 처리
- 출장이 잦아서, 답변에 시간이 좀 걸릴 수도 있습니다. 양해 바랍니다. - 학습 관련 질문을 남겨주세요. 상세히 작성하면 더 좋아요! - 먼저 유사한 질문이 있었는지 검색해보세요. - 서로 예의를 지키며 존중하는 문화를 만들어가요. - 잠깐! 인프런 서비스 운영 관련 문의는 1:1 문의하기를 이용해주세요. print를해서 error를 처리했다고 표현하셨는데 해결을 했다는 의미가 아니고 그냥 print한 것이 처리했다는 의미가 맞나용? print를 했으니깐 reset해주는 것이고요 PeekAtLastError()는 print도 안하고 하면 볼 수 있는 것도 없는건가요?
-
미해결CUDA 프로그래밍 (1) - C/C++/GPU 병렬 컴퓨팅 - CUDA 커널 kernel
CUDA Samples 없음
- 출장이 잦아서, 답변에 시간이 좀 걸릴 수도 있습니다. 양해 바랍니다. v12.8인데 samples 파일이 없습니다.- 학습 관련 질문을 남겨주세요. 상세히 작성하면 더 좋아요! - 먼저 유사한 질문이 있었는지 검색해보세요. - 서로 예의를 지키며 존중하는 문화를 만들어가요. - 잠깐! 인프런 서비스 운영 관련 문의는 1:1 문의하기를 이용해주세요.
-
미해결CUDA 프로그래밍 (4) - C/C++/GPU 병렬 컴퓨팅 - 행렬 matrix 곱하기
24-2 3중 for loop 개선책 이해가 가지 않습니다.
- 출장이 잦아서, 답변에 시간이 좀 걸릴 수도 있습니다. 양해 바랍니다. - 학습 관련 질문을 남겨주세요. 상세히 작성하면 더 좋아요! - 먼저 유사한 질문이 있었는지 검색해보세요. - 서로 예의를 지키며 존중하는 문화를 만들어가요. - 잠깐! 인프런 서비스 운영 관련 문의는 1:1 문의하기를 이용해주세요. 안녕하세요 좋은 강의 감사합니다.3중 for loop 개선책 설명해주신 부분 이해가 가지 않아 질문드립니다.말씀하신건 A도 B도 C도 캐쉬 친화적으로 이동 한다고 말씀하셨는데 A배열은 아래로 움직이지 않나요? 물론 matmul-host.cpp 보다는 캐쉬 미스 확률이 좋아지겠지만 A배열의 인덱스 이동은 캐쉬 친화적인거 같지 않은데 강의에서는 A도 옆으로 움직인다고 하셔서 헷갈려서 문의드립니다.
-
미해결CUDA 프로그래밍 (4) - C/C++/GPU 병렬 컴퓨팅 - 행렬 matrix 곱하기
transpose-shared.cu 가 느린 이유에 대해서 질문 있습니다.
- 출장이 잦아서, 답변에 시간이 좀 걸릴 수도 있습니다. 양해 바랍니다. - 학습 관련 질문을 남겨주세요. 상세히 작성하면 더 좋아요! - 먼저 유사한 질문이 있었는지 검색해보세요. - 서로 예의를 지키며 존중하는 문화를 만들어가요. - 잠깐! 인프런 서비스 운영 관련 문의는 1:1 문의하기를 이용해주세요. 안녕하세요, 좋은 강의 감사합니다.transpose-shared.cu가 느린 이유에 대해 질문이 있습니다.강의 23-2에서는 X축으로 읽는 것이 아니라 Y축으로 메모리를 사용했기 때문에 느려졌다고 설명해 주셨습니다.보통 이런 경우, 이중 for문을 돌릴 때 메모리 공간 지역성(spatial locality)으로 인해 X축(연속된 메모리)으로 읽는 것보다 Y축(연속되지 않은 메모리)으로 읽는 것이 더 느리다고 알고 있습니다.하지만 이 transpose-shared.cu CUDA 예제에서는 각 스레드가 배열의 단 하나의 인덱스만 사용하므로,이중 for문에서 발생하는 문제라기보다는 이후 강의에서 설명해 주신 Bank Conflict가 발생해서 느려진 것이 아닐까 생각했습니다.제 생각이 맞는지 여쭤보고 싶습니다. 수정 및 추가 질문)챕터16강의 다시 보고 제 생각이 틀렸다는것을 인지했습니다.그럼 결국 데이터를 저장할때도 캐쉬 문제인것일까요?transpose-block.cu 는 캐쉬에 저장하고 한번에 flush가 가능하지만transpose-shared.cu 는 저장할때마다 캐쉬미스가 발생해서 매번 global memory에 접근해야하니 느려지는것일까요?
-
미해결CUDA 프로그래밍 (3) - C/C++/GPU 병렬 컴퓨팅 - 메모리 구조
pitch값에 따른 alignement boundary(size) 질문
안녕하세요, 선생님의 좋은 강의 덕분에 잘 공부하고 있습니다.강의 내용 복기 중에 한가지 의문이 생겨 질문 남깁니다.강의 17-8 matrix addition, pitched의 4분 07초 부분에서 device pitch (dev_pitch) 값이 40448 byte로 나온 것을 확인할 수 있고, 이어서 이 경우에 CUDA GPU 메모리 alignment boundary가 256 byte라고 설명 주셨습니다.여기서, 만약 alignment boundary가 256 byte 라면 device pitch 값은 40448 byte가 아닌 40192 byte 가 나와야 하지 않는지 의문이 들었습니다. 40192 byte가 40000 byte(10000개 elements) 에서 가장 가까운 256의 배수임과 동시에 40000 byte를 모두 커버할 수 있다는 근거로 이렇게 생각을 했습니다.그래서 alignment boundary 값을 512 byte로 계산을 해보니, 40448 byte가 40000 byte로부터 가장 가까운 512 배수로 계산이 되는데요. 이 경우에, 어떤 값이 맞는지 확신이 서지 않아 질문 글을 작성하게 되었습니다.제가 사용한 계산식은 pitch = ceil(row_size_in_byte / alignment_boundary_in_byte) * alignment_boundary_in_byte 입니다. 제가 아직 공부하는 단계이고, 이해가 부족했을 가능성이 큽니다. 답변에 소중한 시간 내어주시면 정말 감사드립니다. 다시 한번 멋진 강의 만들어주셔서 감사합니다.
-
미해결CUDA 프로그래밍 (2) - C/C++/GPU 병렬 컴퓨팅 - 벡터 vector 더하기
common.cpp의 procArg 함수에 대해 질문이 있습니다.
- 출장이 잦아서, 답변에 시간이 좀 걸릴 수도 있습니다. 양해 바랍니다. - 학습 관련 질문을 남겨주세요. 상세히 작성하면 더 좋아요! - 먼저 유사한 질문이 있었는지 검색해보세요. - 서로 예의를 지키며 존중하는 문화를 만들어가요. - 잠깐! 인프런 서비스 운영 관련 문의는 1:1 문의하기를 이용해주세요.안녕하세요! C++과 CUDA 공부를 쉽고 재미있게 잘 하고 있습니다 ^^ 부족해서 하나부터 다 공부하고 있는데요.중요한 건 아닌 것 같습니다만 common.cpp의 procArg 함수의 이하 if문의 conditoin에 대한 질문이 생겨서 글을 남기게 되었습니다.if (typeid(TYPE) == typeid(float) && typeid(TYPE) == typeid(double)) 다름이 아니라 TYPE의 변수 타입을 알아내 strtof인지 strtol을 사용하는 결정의 if문의 조건에 걸린 연산자가 || (or) 연산자여야 할 것 같은데요..어차피 vecSize의 개수는 정수로 떨어지니 저렇게 만들어두셨나 싶다가도 정 궁금해서 여쭤봅니다! 그럼 답변 기다리겠습니다.감사합니다 😀
-
미해결CUDA 프로그래밍 (5) - C/C++/GPU 병렬 컴퓨팅 - 아토믹 연산 atomic op
[30-4] hist-shared.cu 에서 질문있습니다.
"hist-shared.cu"의 다음 부분에서 질문있습니다.if(threadIdx.x < HIST_SIZE) { atomicAdd(&(hist[threadIdx.x]), s_hist[threadIdx.x]); }threadIdx.x는 0~31로 모두 다르기 때문에 병렬 연산이긴 해도, global memory에 있는 hist 배열의 서로 다른부분으로 write가 이루어진다고 생각했습니다. 따라서 서로 겹치는 위치가 없어서 atomic한 연산이 필요없을거라 생각했는데, atomic 연산이 없으면 결과가 이상하게 나오더라구요.. 혹시 왜 그런지 알 수 있을까요? 양질의 강의 제공해주셔서 감사합니다!
-
해결됨CUDA 프로그래밍 (3) - C/C++/GPU 병렬 컴퓨팅 - 메모리 구조
[Matrix Addition Problem] 1D를 사용하지 않는 이유가 뭔가요?
강의에서 10,000 x 10,000 행렬의 덧셈을 해결하고자 했고, 이를 2D thread block(32 x 32)으로 풀어내셨습니다. 1D thread block을 사용하지 않는 것에 대한 의문이 들었습니다. [질문]1D thread block은 실무에서 잘 쓰이지 않나요? 의문을 가진 이유는 다음과 같습니다. 우선 2D로 설계할 경우, thread block의 dimension을 어떻게 설정하는지에 따라 낭비되는 thread가 달라집니다.[1 x 1,024] : GridDim(10,000 x 10 x 1), 2,400,000개 낭비[2 x 512] : GridDim(5,000 x 20 x 1), 2,400,000개 낭비[4 x 256] : GridDim(2,500 x 40 x 1), 2,400,000개 낭비[8 x 128] : GridDim(1,250 x 79 x 1), 1,120,000개 낭비[16 x 64] : GridDim(625 x 157 x 1), 480,000개 낭비[32 x 32] : GridDim(313 x 313 x 1), 320,256개 낭비 Thread 낭비가 가장 적은 block dimension은 강의에서 소개해주신 [32 x 32] 입니다. 저는 여기서 1D로 설계했을 때, 낭비되는 therad의 개수가 궁금해졌습니다.[1,024] : GridDim(97,657 x 1 x 1), 768개 낭비 1D로 thread block을 설계했을때, therad 낭비는 768개로 2D로 설계하는 경우와 비교했을때 적지않은 차이라고 생각합니다. 물론 실제로 측정해본 결과(RTX 3090) 약 40 usec 차이로 1D thread block인 경우가 더 빠르긴 했지만 생각보다 큰 차이는 아니었습니다.양질의 강의를 제공해주셔서 감사합니다!!
-
해결됨CUDA 프로그래밍 (2) - C/C++/GPU 병렬 컴퓨팅 - 벡터 vector 더하기
Thread Block 개수에 제한이 있나요?
안녕하세요 저번 Vector Addition 수업에서는 256M(256 x 1024 x 1024)개의 연산을 수행하기 위해서<<<256*1024, 1024>>>로 커널을 런치하여 Block의 개수에는 제한이 없는 것 처럼 보였는데,15-5강의에서는 SM의 Thread Block은 32개 까지 가능이라고 하셔서 헷갈리는데두 개의 Thread Block 개념이 다른 것인지 궁금합니다.
-
해결됨CUDA 프로그래밍 (2) - C/C++/GPU 병렬 컴퓨팅 - 벡터 vector 더하기
Warp를 동시에 돌린다는 말이 이해가 가지 않습니다.
안녕하세요. 먼저 좋은 CUDA 강의를 제공해주셔서 감사합니다. 다름이 아니라 15-4 강의에서 Warp를 동시에 돌린다는 말에 질문이 있습니다. Warp 스케줄링을 저는 아래와 같이 비교하여 이해하였습니다.CPU: Core -> 여러개의 쓰레드GPU: SM -> 여러개의 Warp 강의 내용처럼, CUDA에서 Warp별로 스케줄링을 관리를 한다면 모든 SP의 Clock들이 같은 Instruction을 수행(SM내 SP들의 Clock들은 Sync가 맞음)하여야 Warp단위로 스케줄링이 가능할 것 같은데, 제 생각이 맞는지 궁금합니다. 감사합니다.
-
미해결GPU 프로그래밍 언어 CUDA(쿠다) 기초
ppt 자료는 없나요..
ppt 자료는 없나요..수업중에 ppt에 포함되어 있다고 말씀하셨는데.. ppt 자료는 어디서 구할 수 있나요
-
해결됨CUDA 프로그래밍 (3) - C/C++/GPU 병렬 컴퓨팅 - 메모리 구조
계산시간 비교
알찬 강의 계속 잘 듣고 있습니다!CPU와 GPU version의 계산시간 비교에 대해 질문 드립니다. 커널에서 계산시간만 비교하면 CUDA이 훨씬 빠릅니다. CPU version (filter-host.cu): 44,784 usCUDA version (filter-dev.cu): 830 us하지만 CUDA version에서 메모리 복사를 포함하면 337,737 us 으로 CPU version보다 훨씬 느린 것 같습니다. 혹시 메모리 복사 시간을 줄이는 방법이 있는지요?
-
해결됨CUDA 프로그래밍 (2) - C/C++/GPU 병렬 컴퓨팅 - 벡터 vector 더하기
매크로 함수를 쓴 이유
안녕하세요!좋은 강의 잘 듣고 있습니다.common.cpp 코드 중에서 ELAPSED_TIME_BEGIN(N)과 ELAPSED_TIME_BEGIN(N)을 매크로 함수로 정의한 이유가 무었인지요?CUDA_CHECK_ERROR() 함수는 __FILE__과 __LINE__ 때문에 매크로 함수로 정의하신 걸로 이해합니다만, 시간 측정 함수들을 매크로함수로 정의하신 이유는 잘 이해되지 않습니다.답변 부탁드립니다.감사합니다.
-
해결됨CUDA 프로그래밍 (1) - C/C++/GPU 병렬 컴퓨팅 - CUDA 커널 kernel
hello-prrallel.cu 의 병렬처리에 관하여
좋은 강의 고맙습니다.병렬처리라면 일반적으로 다른 쓰레드에 영향을 주지 않고 수행이 될거라 생각이 됩니다.예제 코드에서 hello<<<1,8>>>();을 이용하여 쓰레드 8개를 병렬로 돌리는데 출력된 값은hello CUDA 0!hello CUDA 1!hello CUDA 2!,,,hello CUDA 7!이 되는 것을 볼 수 있는데 병렬로 처리된다면 출력 되는 숫자가 순차적으로 나올 수 없는 것 아닌가요?
-
해결됨CUDA 프로그래밍 (4) - C/C++/GPU 병렬 컴퓨팅 - 행렬 matrix 곱하기
kernelMatCpy에서는 __syncthreads가 필요없지 않나요
제목 그대로 kernelMatCpy에서는 각 스레드 작업이 다른 스레드에 영향을 주지 않기 때문에 필요없을 것 같은데요. 실제로 __stncthreads()를 제거하고 돌려봐도 정상적으로 결과가 나오구요.혹시 이런 경우라도 내부적으로 꼬일 수 있어서 사용하신 것인지 아니면 대부분의 shared memory로 복사해서 쓰는 경우에 필요하기 때문에 습관(?)차원에서 사용하신 것인지 궁금합니다.
-
해결됨CUDA 프로그래밍 (4) - C/C++/GPU 병렬 컴퓨팅 - 행렬 matrix 곱하기
소스 코드 에러 문의 드립니다
소스코드 그대로 돌리면, 아래와 같은 에러가 발생합니다. (25장 전체 소스가 모두 동일합니다) 이유와 수정 방법을 알고 싶읍니다. // 25d-gemm-alignedTile.cu/tmp/tmpg97edtlu/25d-gemm-alignedTile.cu(98): error: "beta" is ambiguous 1 error detected in the compilation of "/tmp/25d-gemm-alignedTile.cu".