CUDA 프로그램의 제어 흐름

Question

아래는 CUDA에서 경쟁 조건이 어떻게 발생 하는지를 알기 위해 작성한 작은 프로그램이지만 출력에 놀랐습니다.

#include<cutil.h> 
#include<iostream> 
__global__ void testLocal(int *something, int val[]){ 

*something = *something/2; 


val[threadIdx.x] = *something; 
} 

void main(){ 

    int *a, *c; 
    int r =16; 

    cudaMalloc((void**)&a, 4*sizeof(int)); 
    cudaMalloc((void**)&c, sizeof(int)); 
    cudaMemcpy(c, &r, sizeof(int) , cudaMemcpyHostToDevice); 
    testLocal<<<1,4>>>(c,a); 
    int *b = (int *)malloc(4 * sizeof(int)); 
    cudaMemcpy(b,a, 4 * sizeof(int), cudaMemcpyDeviceToHost); 

    for(int j =0 ; j< 4; j++){ 
     printf("%d\n",b[j]); 

    } 
    getchar(); 


} 

나는 4 개의 스레드를 실행하기 때문에 각 스레드가 * 2로 1을 나누기를 기대했다. 나는 그들이 어떤 것을 나눌 순서가 고정되어 있지 않다는 것을 이해한다. 따라서 값을 인쇄하려 할 때 인쇄 된 값 중 하나는 8이고 하나는 4, 하나는 2, 하나는 1이 될 것으로 예상했습니다. 그러나 인쇄 된 값은 모두 8이었습니다. 왜 이것이 ? 모든 스레드가 * 무언가를 한 번 분할해서는 안됩니다.

Answer 1

보고있는 내용은 정의되지 않은 동작입니다. 4 개의 스레드로 단일 블록을 시작하기 때문에 모든 스레드가 동일한 워프에서 실행됩니다. 즉, 실행 한 모든 스레드가 동시에 실행 중임을 의미합니다.

*something = *something/2; 

CUDA 프로그래밍 모델은 동일한 워프의 여러 스레드가 동일한 메모리 위치에 쓰기를 시도 할 때 쓰기 중 하나가 성공한다는 것을 보장합니다. 어떤 스레드가 성공할 것인가와 워프에서 "이기지"않는 다른 스레드에 어떤 일이 생길지에 대해서는 아무 것도 말하지 않습니다. 예상되는 동작을 얻기 위해서는 일련 화 된 메모리 액세스가 필요합니다. 이는이를 지원하는 아키텍처에서 원자 적 메모리 액세스 프리미티브를 사용하는 경우에만 가능합니다.

Answer 2

은 강력한 단어입니다. 당신이하고있는 일이 불특정이기 때문에 은 일이 아니어야합니다.

자,이 가능성이 않는 것은 같은 워프 내에서, 같은 컴퓨팅 장치의 4 개 스레드를 실행됩니다. ("SIMT"모델은 각 스레드가 워프의 일부로 실행되도록합니다). something에서의 작업은 원자가 아니므로 워프 내의 모든 스레드는 잠금 단계에서 메모리를 읽고 씁니다. 따라서 4 개의 스레드는 *something을 함께 읽은 다음 결과를 모두 2로 나눕니다. 모두 8을 메모리에 씁니다. 거기에는 원자 분열 없거나 CUDA에서 사용할 곱 불구하고

당신이 기대했던 어떤

, 그 *something을 읽고 원자 원자 작업을 통해 이루어진다 를 기록. 따라서 이 실제로 인 경우 원한다면 직접 작성해야합니다 (atomicCAS의 도움으로). 또한 병렬로 실행하기 어려운 스레드를 연속적으로 실행해야하므로 성능이 크게 떨어집니다.