GPU CUDA планирование внутренней резьба не работает должным образом

Question

пытается сделать следующий код работать

__global__ void kernel(){ 
    if (threadIdx.x == 1){ 
      while(var == 0){ 
      } 
    } 
    if (threadIdx.x == 0){ 
      var = 1; 
    } 
} 

где вар является переменным глобальным устройством. Я просто запускаю два потока в том же блоке с помощью ядра < < < 1,2 >>>();

Если я переключу порядок ifs, код завершается. Однако, , если я не переключаю порядок ifs, код не заканчивается. Кажется, что если один поток переходит в бесконечный цикл, то никакому другому потоку не будет выделено время выполнения до этого потока завершает весь его код.

У меня создалось впечатление, что в графическом процессоре все потоки получают определенное время от времени (хотя заказ может быть нам неизвестен).

Я также попытался положить __threadfence() внутри цикла while и внутри операторы ifs, а также попытался поместить в цикл while printf. Это все еще не работает.

Что происходит? Любая обратная связь будет оценена.

Спасибо!

Answer 1

Если var - это своего рода глобальная переменная, то, что вы видите, имеет смысл, когда вы рассматриваете, как планируются инструкции из потоков. Вам нужно пройти через свой код, поскольку вы являетесь основой нитей (32 потока). Дивергенция - это когда некоторые из этих 32 потоков выполняют некоторый код, а другие - нет. Когда происходит расхождение, только потоки, выполняющие одну и ту же команду, фактически выполняются до тех пор, пока другие потоки не поймают резервную копию.

Другими словами ...

__global__ void kernel(){ 
    //Both threads encounter this at the same time. Thread 0 is set on "hold" while thread 1 continues in the if block. 
    if (threadIdx.x == 1){ 
       while(var == 0){ 
       }//infinite loop, Thread 0 will always be on hold. Thread 1 will always be in this loop 
     } 

     if (threadIdx.x == 0){ 
       var = 1; 
    } 
} 

в противоположность ...

__global__ void kernel(){ 
    //Both threads encounter this at the same time. Thread 1 is set on "hold" while thread 0 continues in the if block. 
     if (threadIdx.x == 0){ 
     //thread 1 sets global variable var to 1 
       var = 1; 
    } 
    //Threads 1 and 0 join again. 
    //Both encounter this. Thread 0 is set on hold while thread 1 continues. 
    if (threadIdx.x == 1){ 
     //var was set to 1, this is ignored. 
       while(var == 0){ 
       } 
     } 
    //Both threads join 

} 

Пересмотреть руководство по программированию и обзор перекосов. Если вы хотите проверить это дальше, попробуйте поместить оба потока в два блока, это не позволит им находиться в одном и том же варпе.

Будьте предупреждены, хотя CUDA в целом не гарантирует порядок выполнения нитей между перекосами и блоками (если не используется синхронизация какого-либо метода __syncthreads() или выход из ядра).