Как выполнить реляционное соединение на двух контейнерах данных на графическом процессоре (предпочтительно CUDA)?

Question

Что я пытаюсь сделать:

На GPU, я пытаюсь имитировать соглашения, используемые SQL в реляционной алгебре для выполнения объединений в таблицах (например, Inner Join, Outer Join, Cross Join). В приведенном ниже коде я хочу выполнить Inner Join. Представьте себе две таблицы (контейнеры), где одна таблица является родительской/основной таблицей, а другая - дочерней таблицей. Отношения между родителями и дочерними элементами от 1 до многих (или 1 к одному, если нет дочерних элементов Child_ParentID, которые соответствуют элементу в Parent_ID).

Пример входных данных:

Parent_IDs: [1, 2, 3, 4, 5] ... 5 elements 
Parent_Values: [0, 21, 73, 0, 91] ... 5 elements 
Child_ParentIDs: [1, 1, 1, 2, 3, 5, 5] ... 7 elements 
Child_Permanences: [120, 477, 42, 106, 143, 53, 83] ... 7 elements 
Child_Values:  [0, 0, 0, 0, 0, 0, 0] ... 7 elements 

Операция в качестве запроса SQL: Описание

SELECT child.permanence * parent.value FROM child, parent WHERE child.parent_id = parent.id; 

Операция:

Регистрация Child_ParentIDs чтобы получить доступ к Parent_IDs соответствующие Parent_Values. Используйте соответствующие Parent_Values ​​для умножения на соответствующие Child_Permanences и поместите результат каждой операции в Child_Values.

Ожидаемый результат (Child_Values ​​единственный измененный вектор во время операции):

Child_ParentIDs: [1, 1, 1, 2, 3,  5, 5]  ... 7 elements 
Child_Permanences: [120, 477, 42, 106, 143, 53, 83] ... 7 elements 
Child_Values:  [0, 0, 0, 2226, 10439, 4823, 7553] ... 7 elements 

Объяснение (в случае, если это не имеет смысла):

Значение 2226 получается путем умножения 106 и 21. 10439 было умножено на 143 и 73. Также обратите внимание, что ВСЕ записи сохраняются на дочерних векторах (все 7 элементов все еще существуют в выходе, хотя с отдельными элементами Child_Values ​​обновлены). Родительские векторы не сохраняются на выходе (уведомление ParentID 4 отсутствует в списке векторов, и там нет «фиктивного» заполнителя). Это поведение «Внутренняя связь».

Идеи элегантных решений, которые я не получил работу:

-Utilizing CUDA Динамический параллелизм. Возможно, единственное решение для всего Интернета, которое я нашел, делает именно то, что я пытаюсь сделать, это here-part 1 и here-part 2.

-использование хеширующих операций CUDPP;

-Alenka DB.

И, наконец, мой вопрос подтвердил:

Есть ли рабочий раствор с чисто GPU перспектив (желательно с CUDA, но OpenCL будет работать тоже) для выполнения Relational Присоединяется на два отдельных контейнерах данных, так что данные можно выполнять поиск, а элементы обновляются параллельно через указанные соединения?

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

Answer 1

Это похоже на простое умножение элементов между элементами Child_Permanences и выделенные элементы от Parent_Values. С помощью нескольких уставов это можно сделать с помощью одного thrust::transform.

thrust::transform(
    Child_Permanences.begin(), 
    Child_Permanences.end(), 
    thrust::make_permutation_iterator(
     Parent_Values.begin(), 
     thrust::make_transform_iterator(Child_ParentIDs.begin(), 
             _1 - 1)), 
    Child_Values.begin(), 
    _1 * _2); 

Вы можете заметить, что Parent_IDs не используется. Это ограничение приведенного выше кода. Код предполагает, что Parent_IDs может быть не чем иным, как 1-основанием. Вы обнаружите, что thrust::make_transform_iterator не требуется, если Parent_IDs является 0-базовой последовательностью, или Child_ParentIDs является только индексом родительского значения, как указано ниже.

Child_ParentIDs: [0, 0, 0, 1, 2, 4, 4] 

EDIT

Приведенный выше код предполагает, что 1) нет осиротевших детей; и 2) Parent_IDs фиксированный 1 на основе последовательности, как 1, 2, 3, ...

---

при условии, что

1. нет никакого осиротевших детей;
2. Parent_IDs неупорядоченный и уникальный;
3. Child_ParentIDs не распадается, но не является уникальным;

и тот факт, что ваш Parent_IDs имеет тип int16, вы можете создать индексную таблицу родительского значения для дочернего элемента, чтобы посмотреть, когда диапазон Parent_IDs является достаточно малым.

Предполагая, что диапазон Parent_IDs равно [1, 32767], код решением может быть

thrust::device_vector<int> Parent_index(32768, -1); 
thrust::scatter(thrust::make_counting_iterator(0), 
       thrust::make_counting_iterator(0) + Parent_IDs.size(), 
       Parent_IDs.begin(), 
       Parent_index.begin()); 
thrust::transform(
    Child_Permanences.begin(), 
    Child_Permanences.end(), 
    thrust::make_permutation_iterator(
     Parent_Values.begin(), 
     thrust::make_permutation_iterator(
      Parent_index.begin(), 
      Child_ParentIDs.begin())), 
    Child_Values.begin(), _1 * _2); 

Обратите внимание, что Parent_index должны быть созданы заново каждый раз, когда родительский вектор модифицирован.