Поток управления для программы CUDA - PullRequest
Новая
       19

Поток управления для программы CUDA

2 голосов
/ 15 августа 2011

Ниже приведена небольшая программа, которую я написал, чтобы увидеть, как могут возникнуть условия гонки в 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 раза. Я понимаю, что порядок, в котором они будут делиться *, не является фиксированным. Таким образом, когда я пытался напечатать значения, я ожидал, что одно из напечатанных значений будет 8, одно будет 4, одно будет 2 и одно будет 1. Однако все напечатанные значения были 8. Почему это ? Не должны ли все потоки разделить * что-то один раз.

Ответы [ 2 ]

1 голос
/ 15 августа 2011

Если - сильное слово.То, что вы делаете, не определено, поэтому не должно ничего особенного не делать.

Теперь то, что вероятно делает, это запускает 4 потока на одном вычислительном блокев пределах того же основы .(модель "SIMT" заставляет каждый поток работать как часть деформации).Поскольку ваша операция с something не является атомарной, все потоки в оперативной памяти для чтения и записи в режиме блокировки.Таким образом, 4 потока читают *something вместе, затем все делят результат на 2, и все пытаются записать в память 8.

Что вы ожидали, что *something будет прочитано и записано атомарно достигается посредством атомарных операций, хотя в CUDA нет атомного деления или умножения.Так что если вы действительно хотите этого, вам нужно написать свое собственное (с помощью atomicCAS).И вы начнете видеть, как резко падает ваша производительность, поскольку теперь вы заставляете потоки, которые изо всех сил стараются работать параллельно, работать последовательно.

1 голос
/ 15 августа 2011

То, на что вы смотрите - это неопределенное поведение. Поскольку вы запускаете один блок с 4 потоками, все потоки выполняются в одной и той же деформации. Это означает, что 

 *something = *something/2;

выполняется одновременно всеми запущенными вами потоками. Модель программирования CUDA гарантирует только то, что когда несколько потоков из одной и той же деформации пытаются выполнить запись в одну и ту же область памяти, одна из операций записи будет успешной. В нем ничего не говорится о том, какой поток преуспеет, и что произойдет с другими потоками в варпе, которые не «побеждают». Чтобы получить ожидаемое поведение, потребуется сериализованный доступ к памяти - это возможно только при использовании примитивов атомарного доступа к памяти на тех архитектурах, которые их поддерживают.

...