CUDA - память устройства, поиск строки в тексте

Question

У меня есть строка длиной не более 500 символов и текстовый файл размером 200 МБ. Я хочу написать программу в CUDA для поиска строки в текстовом файле. Мой текстовый файл слишком большой, и я думаю, что я должен поместить его в глобальную память устройства, но как насчет моей строки? Что является лучшим среди общей, постоянной и текстурной памяти? и почему? Также у меня есть массив размером максимум 2500. Какие типы памяти устройства подходят для его хранения?

Answer 1

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

Вот некоторые псевдо-код для простой базовой реализации

...load blocksize+strlen bytes of the file into shared memory... 
__syncthreads(); 
bool found = true; 
for (int i = 0; i < strlen; i++) { 
    if (file_chunk_in_sharedmem[threadIdx.x + i] != 
     search_str_in_constantmem[i]) 
    { 
     found = false; 
     break; 
    } 
} 
if (found) { 
    ...output the result... 
} 

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

Функции профилирования и/или cuda - ваши друзья.

Answer 2

Для наивной реализации на Fermi:

Сохраните текстовый файл в глобальной памяти и строка поиска в постоянной памяти. Настройте буфер результатов того же размера, что и текстовый файл. Заполните буфер результатов нулями.

Позвольте числу потоков в каждом блоке, t, быть таким же, как длина строки поиска. Чтобы определить размер сетки, рассмотрите размер вашего текстового файла и ограничение размера сетки в 64 КБ. Чтобы охватить весь ваш файл, выберите размер для x, например, 10K. Затем найдите размер для y, разделив размер вашего текстового файла на x и округляя результат. Таким образом, 200M/10K = 20K (который находится в пределах 64K). Запустить ядро ​​с помощью t резьба и (x, y) сетка.

В ядре:

вычислить смещение в текстовый файл, как д = х + 1024 * у.

Поскольку размер y был округлен сверху, некоторые ядра в конце пробега должны быть прерваны. Отменить нить, если d + t больше, чем размер текстового файла.

Else, имеют нагрузки нить один символ с индексом т из строки поиска и сравнить его с одним символом с индексом т + д в текстовом файле. Если символы не совпадают, сохраните «1» в буфере результатов с индексом d, иначе ничего не сделайте.

Когда ядро ​​завершает работу, просмотрите буфер результатов с помощью Thrust. Каждое место, которое равно 0, обозначает начальную точку матча.