sorting - CUDA Thrust y sort_by_key

permutation

Question

Estoy buscando un algoritmo de clasificación en CUDA que pueda ordenar una matriz A de elementos (doble) y devuelva una matriz de claves B para esa matriz A. Conozco la función sort_by_key en la biblioteca Thrust pero quiero mi matriz de elementos A permanecer sin cambios ¿Que puedo hacer?

Mi código es:

void sortCUDA(double V[], int P[], int N) {
real_t *Vcpy = (double*) malloc(N*sizeof(double));
memcpy(Vcpy,V,N*sizeof(double));
thrust::sort_by_key(V, V + N, P);
free(Vcpy);
}

Estoy comparando el algoritmo de empuje con otros que tengo en la CPU secuencial

N mergesort sortCUDA
113 0.000008 0.000010
226 0.000018 0.000016
452 0.000036 0.000020
905 0.000061 0.000034
1810 0.000135 0.000071
3621 0.000297 0.000156
7242 0.000917 0.000338
14484 0.001421 0.000853
28968 0.003069 0.001931
57937 0.006666 0.003939
115874 0.014435 0.008025
231749 0.031059 0.016718
463499 0.067407 0.039848
926999 0.148170 0.118003
1853998 0.329005 0.260837
3707996 0.731768 0.544357
7415992 1.638445 1.073755
14831984 3.668039 2.150179
115035495 39.276560 19.812200
230070990 87.750377 39.762915
460141980 200.940501 74.605219

El rendimiento de empuje no es malo, pero creo que si uso OMP probablemente pueda obtener fácilmente un mejor tiempo de CPU

Creo que esto es porque memcpy

SOLUCIÓN:

void thrustSort(double V[], int P[], int N)
{
thrust::device_vector<int> d_P(N);
thrust::device_vector<double> d_V(V, V + N);
thrust::sequence(d_P.begin(), d_P.end());
thrust::sort_by_key(d_V.begin(), d_V.end(), d_P.begin());
thrust::copy(d_P.begin(),d_P.end(),P);
}

donde V es uno de mis valores dobles para ordenar

Answer 1

¿Qué tan grande es esta matriz? La forma más eficiente, en términos de velocidad, es simplemente duplicar el conjunto original antes de ordenar, si la memoria está disponible.

Answer 2

Sobre la base de la respuesta proporcionada por @asm (no pude hacerlo funcionar), este código pareció funcionar para mí y solo ordena las claves. Sin embargo, creo que está limitado al caso en que las claves están en la secuencia 0, 1, 2, 3, 4 ... correspondiente a los valores (dobles). Como se trata de un tipo de "valor de índice", podría extenderse al caso de una secuencia arbitraria de claves, tal vez haciendo una copia indexada. Sin embargo, no estoy seguro de que el proceso de generar la secuencia de índice y reorganizar las claves originales sea más rápido que simplemente copiar los datos de valor original a un nuevo vector (para el caso de claves arbitrarias).

#include <iostream>
#include <thrust/device_vector.h>
#include <thrust/host_vector.h>
#include <thrust/sort.h>
using namespace std;
__device__ double *rawA; // an array in global mem
struct cmp : public binary_function<int, int, bool>
{
__host__ __device__ bool operator()(const int i, const int j) const
{return ( rawA[i] < rawA[j]);}
};
void sortkeys(double *A, int n) {
// move data to the gpu
thrust::device_vector<double> devA(A, A + n);
// rawA = thrust::raw_pointer_cast(&(devA[0]));
double *test = raw_pointer_cast(devA.data());
cudaMemcpyToSymbol(rawA, &test, sizeof(double *));
thrust::device_vector<int> B(n);
// initialize keys
thrust::sequence(B.begin(), B.end());
thrust::sort(B.begin(), B.end(), cmp());
// B now contains the sorted keys
thrust::host_vector<int> hostB = B;
for (int i=0; i<hostB.size(); i++)
std::cout << hostB[i] << " ";
std::cout<<std::endl;
for (int i=0; i<hostB.size(); i++)
std::cout << A[hostB[i]] << " ";
std::cout<<std::endl;
}
int main(){
double C[] = {0.7, 0.3, 0.4, 0.2, 0.6, 1.2, -0.5, 0.5, 0.0, 10.0};
sortkeys(C, 9);
std::cout << std::endl;
return 0;
}

Answer 3

Puede modificar el operador de comparación para ordenar claves en lugar de valores. @Robert Crovella señaló correctamente que un puntero de dispositivo sin procesar no se puede asignar desde el host. El algoritmo modificado está a continuación:

struct cmp : public binary_function<int,int,bool>
{
cmp(const double *ptr) : rawA(ptr) { }
__host__ __device__ bool operator()(const int i, const int j) const
{return rawA[i] > rawA[j];}
const double *rawA; // an array in global mem
};
void sortkeys(double *A, int n) {
// move data to the gpu
thrust::device_vector<double> devA(A, A + n);
double *rawA = thrust::raw_pointer_cast(devA.data());
thrust::device_vector<int> B(n);
// initialize keys
thrust::sequence(B.begin(), B.end());
thrust::sort(B.begin(), B.end(), cmp(rawA));
// B now contains the sorted keys
}

Y aquí hay una alternativa con arrayfire. Aunque no estoy seguro de cuál es más eficiente ya que la solución arrayfire usa dos matrices adicionales:

void sortkeys(double *A, int n) {
af::array devA(n, A, af::afHost);
af::array vals, indices;
// sort and populate vals/indices arrays
af::sort(vals, indices, devA);
std::cout << devA << "/n" << indices << "/n";
}