Код CUDA запускается при компиляции с sm_35, но не работает с sm_30

Устройство с графическим процессором, которое у меня есть, — это GeForce GT 750M, которое, как я обнаружил, имеет вычислительную мощность 3.0. Я скачал код CUDA, найденный здесь: (https://github.com/fengChenHPC/word2vec_cbow. Его make-файл имел флаг -arch=sm_35.

Поскольку мое устройство имеет вычислительную мощность 3.0, я изменил флаг на -arch=sm_30. Он скомпилирован нормально, но когда я запускаю код, я получаю следующую ошибку:

word2vec.cu 449 : unspecified launch failure

word2vec.cu 449 : unspecified launch failure

Он показывает это несколько раз, потому что есть несколько потоков ЦП, запускающих ядро ​​CUDA. Обратите внимание, что потоки не используют разные потоки для запуска ядра, поэтому все запуски ядра выполняются по порядку.

Теперь, когда я оставляю флаг, то есть -arch=sm_35, код работает нормально. Может кто-нибудь объяснить, почему код не запускается, когда я устанавливаю флаг, соответствующий моему устройству?


person user1274878    schedule 15.07.2015    source источник
comment
Вы пытались запустить код с помощью cuda-memcheck?   -  person talonmies    schedule 15.07.2015
comment
Итак, запуск с помощью cuda-memcheck показывает множество ошибок нелегального доступа. Есть ли способ добраться до номера строки? Он показывает адрес инструкции, но не номер строки. Но мне до сих пор не ясно, почему он работает нормально с одним вариантом компиляции, но не запускается с другим. Пожалуйста, объясни   -  person user1274878    schedule 15.07.2015
comment
Вы можете создать код с опцией -lineinfo, чтобы заставить cuda-memcheck сообщать номера строк. Я не могу объяснить, не видя кода, и ваша работа — найти код. Я не буду анализировать для вас 1000+ строк кода, извините.   -  person talonmies    schedule 15.07.2015
comment
Я на 100% уверен, что ваше утверждение о том, что код запускается при компиляции для sm_35, совершенно неверно, и что ни одно ядро ​​никогда не запускается при сборке этого кода для вычислительных возможностей 3.5.   -  person talonmies    schedule 16.07.2015


Ответы (2)


К сожалению, ваш вывод о том, что код работает при компиляции для sm_35 и запуске на sm_30 GPU, неверен. Виновник вот в чем:

void cbow_cuda(long window, long negative, float alpha, long sentence_length, 
               int *sen, long layer1_size, float *syn0, long hs, float *syn1, 
               float *expTable, int *vocab_codelen, char *vocab_code,
               int *vocab_point, int *table, long table_size, 
               long vocab_size, float *syn1neg){
    int blockSize = 256;
    int gridSize = (sentence_length)/(blockSize/32);
    size_t smsize = (blockSize/32)*(2*layer1_size+3)*sizeof(float);
//printf("sm size is %d\n", smsize);
//fflush(stdout);
    cbow_kernel<1><<<gridSize, blockSize, smsize>>>
                   (window, negative, alpha, sentence_length, sen,
                    layer1_size, syn0, syn1, expTable, vocab_codelen,
                    vocab_code, vocab_point, table, table_size,
                    vocab_size, syn1neg);
}

Этот код автоматически завершится ошибкой, если запуск ядра завершится ошибкой из-за неполной проверки ошибок API. И запуск ядра невозможен, если вы собираете для sm_35 и запускаете на sm_30. Если вы измените код этой функции на этот (добавив проверку ошибок запуска ядра):

void cbow_cuda(long window, long negative, float alpha, long sentence_length, 
               int *sen, long layer1_size, float *syn0, long hs, float *syn1, 
               float *expTable, int *vocab_codelen, char *vocab_code,
               int *vocab_point, int *table, long table_size, 
               long vocab_size, float *syn1neg){
    int blockSize = 256;
    int gridSize = (sentence_length)/(blockSize/32);
    size_t smsize = (blockSize/32)*(2*layer1_size+3)*sizeof(float);
//printf("sm size is %d\n", smsize);
//fflush(stdout);
    cbow_kernel<1><<<gridSize, blockSize, smsize>>>
                   (window, negative, alpha, sentence_length, sen,
                    layer1_size, syn0, syn1, expTable, vocab_codelen,
                    vocab_code, vocab_point, table, table_size,
                    vocab_size, syn1neg);
    checkCUDAError( cudaPeekAtLastError() );
}

и скомпилируйте и запустите его для sm_35, вы должны получить это на устройстве sm_30:

~/cbow/word2vec_cbow$ make
nvcc word2vec.cu -o word2vec -O3 -Xcompiler -march=native -w  -Xptxas="-v" -arch=sm_35 -lineinfo
ptxas info    : 0 bytes gmem
ptxas info    : Compiling entry function '_Z11cbow_kernelILx1EEvllflPKilPVfS3_PKfS1_PKcS1_S1_llS3_' for 'sm_35'
ptxas info    : Function properties for _Z11cbow_kernelILx1EEvllflPKilPVfS3_PKfS1_PKcS1_S1_llS3_
    0 bytes stack frame, 0 bytes spill stores, 0 bytes spill loads
ptxas info    : Used 34 registers, 448 bytes cmem[0], 8 bytes cmem[2]

~/cbow/word2vec_cbow$ ./word2vec -train text8 -output vectors.bin -cbow 1 -size 200 -window 7 -negative 1 -hs 1 -sample 1e-3 -threads 1 -binary 1 -save-vocab voc #> out 2>&1
Starting training using file text8
Vocab size: 71290
Words in train file: 16718843
vocab size = 71290
cbow.cu 114 : invalid device function

т.е. запуск ядра завершился неудачно, так как в полезной нагрузке CUDA cubin вашего приложения не было найдено соответствующего кода устройства. Это также отвечает на ваш более ранний вопрос о том, почему вывод этого кода неверен. Ядро анализа просто никогда не запускается на вашем оборудовании, если оно построено с параметрами по умолчанию.

Если я соберу этот код для sm_30 и запущу его на GTX 670 с 2 ГБ памяти (вычислительная способность 3.0), я получу следующее:

~/cbow/word2vec_cbow$ make
nvcc word2vec.cu -o word2vec -O3 -Xcompiler -march=native -w  -Xptxas="-v" -arch=sm_30 -lineinfo
ptxas info    : 0 bytes gmem
ptxas info    : Compiling entry function '_Z11cbow_kernelILx1EEvllflPKilPVfS3_PKfS1_PKcS1_S1_llS3_' for 'sm_30'
ptxas info    : Function properties for _Z11cbow_kernelILx1EEvllflPKilPVfS3_PKfS1_PKcS1_S1_llS3_
    0 bytes stack frame, 0 bytes spill stores, 0 bytes spill loads
ptxas info    : Used 34 registers, 448 bytes cmem[0], 12 bytes cmem[2]

~/cbow/word2vec_cbow$ ./word2vec -train text8 -output vectors.bin -cbow 1 -size 200 -window 7 -negative 1 -hs 1 -sample 1e-3 -threads 1 -binary 1 -save-vocab voc #> out 2>&1
Starting training using file text8
Vocab size: 71290
Words in train file: 16718843
vocab size = 71290
Alpha: 0.000009  Progress: 100.00%  Words/thread/sec: 1217.45k

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

person Community    schedule 17.07.2015

Как видно из этой ССЫЛКИ, GeForce GTX 750M не существует.

ваш либо:

GeForce GTX 750 Ti

GeForce GTX 750

or

GeForce GT 750M

Если у вас один из первых двух, то ваш GPU основан на Maxwell и имеет Compute Capability = 5.0.

В противном случае ваш графический процессор основан на Kepler и имеет Compute Capability = 3.0.

Если вы не уверены, какой у вас GPU, сначала выясните это, запустив deviceQuery из NVIDIA SAMPLE.

person Iman    schedule 15.07.2015
comment
Извините, это GeForce GT 750M - person user1274878; 15.07.2015
comment
Кстати, это надо сделать так: -arch=compute_30 -code=sm_30, проверьте - person Iman; 15.07.2015
comment
Я почти уверен, что это не имеет никакого отношения к проблеме - person talonmies; 16.07.2015