int sphyraena_transfer_data(sphyraena *s)
{
int r;
#ifdef SPHY_DEBUG
sphyraena_timer_start();
#endif
// copy data from cpu to gpu
r = cudaMemcpy(s->data_gpu, s->data_cpu->d,
s->data_cpu->rows * s->data_cpu->stride,
cudaMemcpyHostToDevice);
#ifdef SPHY_DEBUG
sphyraena_timer_end("cudaMemcpy");
#endif
// check for cudaMemcpy error
if(r != cudaSuccess) {
fprintf(stderr, "Cuda error: %s\n", cudaGetErrorString(r));
return SPHYRAENA_ERR_CUDAMEMCPY;
}
return SPHYRAENA_SUCCESS;
}
int sphyraena_test_case(sphyraena *s, const char *sql,
double *time_native, double *time_gpu, double *time_transfer, int *rows_, int streaming, int include_transfer)
{
int rows = 0;
char *err;
int r;
sphyraena_timer_start();
r = sqlite3_exec(s->db, sql, &test_callback, &rows, &err);
time_native[0] = sphyraena_timer_stop();
if(r != SQLITE_OK) {
eprintf(stderr, "SQL error: %s\n%s\n", err, sql);
sqlite3_free(err);
return 1;
}
sphyraena_timer_start();
if(include_transfer == 1)
sphyraena_transfer_data(s);
r = sphyraena_select(s, sql, streaming);
time_gpu[0] = sphyraena_timer_stop();
if(r != SPHYRAENA_SUCCESS) {
eprintf(stderr, "Sphyraena error: case failed\n%s\n", sql);
return 1;
}
sphyraena_timer_start();
sphyraena_transfer_results(s);
time_transfer[0] = sphyraena_timer_stop();
if(rows != s->results_cpu->rows) {
eprintf(stderr, "Incorrect result: %i cpu rows %i gpu rows\n%s\n",
rows, s->results_cpu->rows, sql);
return 1;
}
rows_[0] = rows;
return SPHYRAENA_SUCCESS;
}
int sphyraena_prepare_data(sphyraena *s, const char* sql_stmt)
{
sqlite3_stmt *stmt;
sphyraena_data *data = s->data_cpu;
// get length of null terminated sql_stmt
int i;
for(i = 0; sql_stmt[i] != '\0'; i++);
// prepare our sqlite3_stmt object
sqlite3_prepare_v2(s->db, sql_stmt, i, &stmt, NULL);
#ifdef SPHY_DEBUG
sphyraena_timer_start();
#endif
// step once to examine the results to determine data stride
int r = sqlite3_step(stmt);
// if a row isn't returned when we step return error
if(r != SQLITE_ROW)
return SPHYRAENA_ERR_STMT;
// get and set number of columns in table
int columns = sqlite3_column_count(stmt);
data->columns = columns;
// offset tracks the offset of each column in bytes from the beginning
// of the row, since not every column is 4 bytes
int offset = 0;
sqlite3_value *val;
// go through the first row of results to get all of our data
// types so we know column offsets, don't save column data yet
//
// for each column get it from the sqlite info
for(i = 0; i < columns; i++) {
val = sqlite3_column_value(stmt, i);
// if column type is integer
if(val->type == SQLITE_INTEGER) {
#ifdef USE_INT64
// SQLite does not differentiate between 32 and 64 bit
// ints in the types data column, this controls whether
// or not we want to cast all of the int values to
// int32 or int64
// set the type in the sphyraena_data struct
data->types[i] = SPHYRAENA_INT64;
// set column offset
data->offsets[i] = offset;
// increment offset
offset += sizeof(i64);
#else
data->types[i] = SPHYRAENA_INT;
data->offsets[i] = offset;
offset += sizeof(int);
#endif
}
else if(val->type == SQLITE_FLOAT) {
#ifdef USE_DOUBLES
// see USE_INT64
data->types[i] = SPHYRAENA_DOUBLE;
data->offsets[i] = offset;
offset += sizeof(double);
#else
data->types[i] = SPHYRAENA_FLOAT;
data->offsets[i] = offset;
offset += sizeof(float);
#endif
}
else {
eprintf(stderr,
"Error: the data type for column %i is not supported\n",
i);
return SPHYRAENA_ERR_TYPE;
}
}
// set row stride to the current offset
int stride = offset;
// round stride to next power of 2
/* stride--;
stride |= stride >> 1;
stride |= stride >> 2;
stride |= stride >> 4;
stride |= stride >> 8;
stride |= stride >> 16;
stride++;*/
s->data_cpu->stride = stride;
int rows = 0;
int last_row = floor((float)s->data_size / (float)stride);
char *d = (char*)&s->data_cpu->d;
void *p;
do {
for(i = 0; i < columns; i++) {
p = d + rows * stride + data->offsets[i];
switch(data->types[i]) {
case SPHYRAENA_INT :
((int*)p)[0] = sqlite3_column_int(stmt, i);
break;
case SPHYRAENA_INT64 :
((i64*)p)[0] = sqlite3_column_int64(stmt, i);
break;
case SPHYRAENA_FLOAT :
((float*)p)[0] = (float)sqlite3_column_double(stmt, i);
break;
case SPHYRAENA_DOUBLE :
((double*)p)[0] = sqlite3_column_double(stmt, i);
break;
default:
eprintf(stderr,
"Error: the data type for column %i is not supported\n",
i);
return SPHYRAENA_ERR_TYPE;
}
}
rows++;
if(rows >= last_row) {
eprintf(stderr, "Warning: selected data too large for gpu data block\n");
break;
}
} while(sqlite3_step(stmt) == SQLITE_ROW);
#ifdef COLUMNROW
// the current format is row-column data format, this block performs a
// translation to column-row format, which takes longer, but may make
// query execution faster because of coalescing
char *temp = (char*)malloc(s->data_size);
int offsets[columns];
int j;
for(i = 0; i < rows; i++) {
for(j = 0; j < columns; j++) {
switch(data->types[j]) {
case SPHYRAENA_INT :
((int*)(temp + data->offsets[j] * rows))[i] =
((int*)(d + i * stride + data->offsets[j]))[0];
break;
case SPHYRAENA_INT64 :
((i64*)(temp + data->offsets[j] * rows))[i] =
((i64*)(d + i * stride + data->offsets[j]))[0];
break;
case SPHYRAENA_FLOAT :
((float*)(d + data->offsets[j] * rows))[i] =
((float*)(d + i * stride + data->offsets[j]))[0];
break;
case SPHYRAENA_DOUBLE :
((double*)(temp + data->offsets[j] * rows))[i] =
((double*)(d + i * stride + data->offsets[j]))[0];
break;
}
}
}
for(i = 0; i < columns; i++)
data->offsets[i] *= rows;
memcpy(d, temp, s->data_size);
free(temp);
#endif
data->rows = rows;
#ifdef SPHY_DEBUG
sphyraena_timer_end("SQLite select");
#endif
// clean up sqlite stmt
sqlite3_finalize(stmt);
return SPHYRAENA_SUCCESS;
}
int main(int argc, char **argv)
{
sqlite3 *db;
sphyraena sphy;
int i, r;
int dbarg = -1;
int loadmemory = 0;
int pinned_memory = 0;
for(i = 1; i < argc; i++) {
if(argv[i][0] == '-')
switch(argv[i][1]) {
case 'd' :
dbarg = i + 1;
printf("Using database %s\n", argv[i+1]);
break;
case 'm' :
loadmemory = 1;
printf("Loading database into memory\n");
break;
case 'p' :
pinned_memory = 1;
printf("Using pinned memory\n");
break;
default :
printhelp(argv);
}
}
if(dbarg == -1) {
printhelp(argv);
exit(1);
}
if(loadmemory) {
sqlite3_open(":memory:", &db);
char sql[256];
sprintf(sql, "ATTACH DATABASE '%s' AS loadfrom", argv[dbarg]);
r = sqlite3_exec(db, sql, NULL, NULL, NULL);
sqlite3_exec(db, "CREATE TABLE 'test' AS SELECT * FROM loadfrom.test", NULL, NULL, NULL);
sqlite3_exec(db, "DETACH loadfrom", NULL, NULL, NULL);
}
else {
r = sqlite3_open(argv[dbarg], &db);
}
if(r) {
eprintf(stderr, "Can't open database: %s\n", sqlite3_errmsg(db));
sqlite3_close(db);
exit(1);
}
sphyraena_init(&sphy, db, DATA_SIZE, RESULTS_SIZE, pinned_memory);
int err = sphyraena_init(&sphy, db, DATA_SIZE, RESULTS_SIZE, pinned_memory);
if(err != SPHYRAENA_SUCCESS) {
eprintf(stderr, "Failing to complete init\n");
sqlite3_close(db);
exit(1);
}
#ifndef RUNTEST
sphyraena_timer_start();
sphyraena_prepare_data(&sphy, "SELECT * FROM test");
const char* sql = "SELECT id, uniformi FROM test WHERE uniformi < 90";
char* err;
sphyraena_timer_start();
r = sqlite3_exec(db, sql, NULL, NULL, &err);
double native = sphyraena_timer_end("native execution");
if(err != NULL) {
eprintf(stderr, "exec error: %s\n", err);
exit(1);
}
sphyraena_timer_start();
sphyraena_transfer_data(&sphy);
sphyraena_select(&sphy, sql, 0);
double vm = sphyraena_timer_end("vm execution");
sphyraena_timer_start();
sphyraena_transfer_results(&sphy);
double results = sphyraena_timer_end("transfer results");
printf("stream\n");
printf("speedup: %fx\n", native / vm);
printf("speedup with results: %fx\n", native / (vm + results));
sphyraena_print_results(&sphy, 40);
#else
sphyraena_test_queries(&sphy);
/* this code tests speedup as a function of stream width
sphyraena_test_sizes(&sphy, 0, 1);
sphy.stream_width = 2;
sphyraena_test_sizes(&sphy, 1, 0);
sphy.stream_width = 4;
sphyraena_test_sizes(&sphy, 1, 0);
sphy.stream_width = 8;
sphyraena_test_sizes(&sphy, 1, 0);*/
#endif
sphyraena_cleanup(&sphy);
sqlite3_close(db);
return 0;
}
