CUDA Добавить строки матрицы - PullRequest
Новая
       21

CUDA Добавить строки матрицы

2 голосов
/ 22 июля 2010

Я пытаюсь сложить строки матрицы 4800x9600 вместе, в результате получается матрица 1x9600.

Я разбил 4800x9600 на 9600 матриц длиной 4800 каждая.Затем я выполняю сокращение на 4800 элементов.

Проблема в том, что это действительно медленно ...

Кто-нибудь получил какие-либо предложения?

В принципе, я пытаюсьдля реализации функции sum (...) MATLAB.

Вот код, который я проверял, работает нормально, просто он очень медленный: 

void reduceRows(Matrix Dresult,Matrix DA)
{
        //split DA into chunks
        Matrix Dchunk;
        Dchunk.h=1;Dchunk.w=DA.h;
        cudaMalloc((void**)&Dchunk.data,Dchunk.h*Dchunk.w*sizeof(float));

        Matrix DcolSum;
        DcolSum.h=1;DcolSum.w=1;
        //cudaMalloc((void**)&DcolSum.data,DcolSum.h*DcolSum.w*sizeof(float));

        int i;
        for(i=0;i<DA.w;i++)   //loop over each column
        {
                //printf("%d ",i);
                cudaMemcpy(Dchunk.data,&DA.data[i*DA.h],DA.h*sizeof(float),cudaMemcpyDeviceToDevice);
                DcolSum.data=&Dresult.data[i];
                reduceTotal(DcolSum,Dchunk);
        }
        cudaFree(Dchunk.data);
}

Матрица определяется как: 

typedef struct{
        long w;
        long h;
        float* data;
}Matrix;

ReduceTotal () просто вызывает стандартное сокращение NVIDIA, суммирует все элементы в Dchunk и помещает ответ в DcolSum.

Я собираюсь сделать все это на CPU, если яне могу найти ответ ...; (

Заранее большое спасибо,

Ответы [ 3 ]

3 голосов
/ 27 июля 2010

Вместо зацикливания на каждом столбце распараллеливайте столбцы. Каждый из 4600 потоков суммирует 9600 записей в своем столбце и помещает сумму в соответствующее место в векторе результатов.

Если вы ищете библиотеку для упрощения работы с Cuda, я настоятельно рекомендую Thrust: http://code.google.com/p/thrust/

Используя Thrust, я бы создал функтор для хранения указателя вашей матрицы в памяти устройства, а затем отобразил его на последовательности индексов столбцов. Оператор () функтора берет индекс, суммирует все в этом столбце матрицы и возвращает сумму. Тогда ваша сумма будет храниться в thrust :: device_vector без каких-либо копий памяти (или даже прямых вызовов CUDA).

Ваш функтор может выглядеть примерно так: 

struct ColumnSumFunctor {
    const Matrix matrix;

    // Make a functor to sum the matrix
    ColumnSumFunctor(const Matrix& matrix);

    // Compute and return the sum of the specified column
    __device__
    int operator()(const int& column) const;
};
1 голос
/ 23 июля 2010

Сокращение - это базовая операция в GPGPU, она должна быть быстрой, и 9600 раз сокращения не должны быть медленными.

Какую видеокарту вы используете?

Я предлагаю вам разбить его на 9600 массивов, каждый раз, когда вы сокращаете массив из 4800 элементов в один результат. Вместо Reduto Total я предлагаю вам использовать CUDPP для выполнения операции сокращения, CUDPP подобен STL для CUDA. Это реализовано с заботой о производительности.

http://code.google.com/p/cudpp/

0 голосов
/ 23 июля 2010

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

Самый наивный способ реализовать это - не объединить память, но она может быть быстрее, чем вы делаете это сейчас.

Как только у вас наивный способ работы, объедините чтение вашей памяти: например, каждый поток в блоке читает 16 последовательных операций с плавающей запятой в общую память, синхронизирует потоки, затем накапливает соответствующие 16 операций с плавающей запятой в регистре, синтезирует, затем повторите

В Computing SDK есть много примеров методов сокращения.

...