static void
test_convert_16 (Fixture *fixture,
gconstpointer unused)
{
gfloat *host_data;
host_data = ufo_buffer_get_host_array (fixture->buffer, NULL);
g_memmove (host_data, fixture->data16, fixture->n_data * sizeof (guint16));
ufo_buffer_convert (fixture->buffer, UFO_BUFFER_DEPTH_16U);
host_data = ufo_buffer_get_host_array (fixture->buffer, NULL);
for (guint i = 0; i < fixture->n_data; i++)
g_assert (host_data[i] == ((gfloat) fixture->data16[i]));
}
static void
ufo_hdf5_reader_read (UfoReader *reader,
UfoBuffer *buffer,
UfoRequisition *requisition,
guint roi_y,
guint roi_height,
guint roi_step)
{
UfoHdf5ReaderPrivate *priv;
gpointer data;
hid_t dst_dataspace_id;
hsize_t dst_dims[2];
priv = UFO_HDF5_READER_GET_PRIVATE (reader);
data = ufo_buffer_get_host_array (buffer, NULL);
hsize_t offset[3] = { priv->current, roi_y, 0 };
hsize_t count[3] = { 1, roi_height, requisition->dims[0] };
dst_dims[0] = roi_height;
dst_dims[1] = requisition->dims[0];
dst_dataspace_id = H5Screate_simple (2, dst_dims, NULL);
H5Sselect_hyperslab (priv->src_dataspace_id, H5S_SELECT_SET, offset, NULL, count, NULL);
H5Dread (priv->dataset_id, H5T_NATIVE_FLOAT, dst_dataspace_id, priv->src_dataspace_id, H5P_DEFAULT, data);
H5Sclose (dst_dataspace_id);
priv->current++;
}
static void
read_64_bit_data (UfoTiffReaderPrivate *priv,
UfoBuffer *buffer,
UfoRequisition *requisition,
guint roi_y,
guint roi_height,
guint roi_step)
{
gdouble *src;
gfloat *dst;
dst = ufo_buffer_get_host_array (buffer, NULL);
src = g_new0 (gdouble, requisition->dims[0]);
for (guint i = roi_y; i < roi_y + roi_height; i += roi_step) {
TIFFReadScanline (priv->tiff, src, i, 0);
for (guint j = 0; j < requisition->dims[0]; j++)
dst[j] = (gfloat) src[j];
dst += requisition->dims[0];
}
g_free (src);
}
static gboolean
ufo_averager_task_generate (UfoTask *task,
UfoBuffer *output,
UfoRequisition *requisition)
{
UfoAveragerTaskPrivate *priv;
gfloat *out_array;
gsize n_pixels;
priv = UFO_AVERAGER_TASK_GET_PRIVATE (UFO_AVERAGER_TASK (task));
if (priv->n_generate == 0)
return FALSE;
out_array = ufo_buffer_get_host_array (output, NULL);
n_pixels = requisition->dims[0] * requisition->dims[1];
if (!priv->is_data_averaged) {
for (gsize i = 0; i < n_pixels; i++)
priv->averaged[i] = out_array[i] / (gfloat) priv->counter;
priv->is_data_averaged = TRUE;
}
g_memmove (out_array, priv->averaged, n_pixels * sizeof (float));
priv->n_generate--;
return TRUE;
}
static gboolean
ufo_null_task_process (UfoTask *task,
UfoBuffer **inputs,
UfoBuffer *output,
UfoRequisition *requisition)
{
UfoNullTaskPrivate *priv;
priv = UFO_NULL_TASK_GET_PRIVATE (task);
if (priv->force_download) {
gfloat *host_array;
host_array = ufo_buffer_get_host_array (inputs[0], NULL);
host_array[0] = 0.0;
}
if (priv->finish) {
UfoGpuNode *gpu;
gpu = UFO_GPU_NODE (ufo_task_node_get_proc_node (UFO_TASK_NODE (task)));
UFO_RESOURCES_CHECK_CLERR (clFinish (ufo_gpu_node_get_cmd_queue (gpu)));
}
return TRUE;
}
/**
* ufo_op_euclidean_distance:
* @arg1: A #UfoBuffer
* @arg2: A #UfoBuffer
* @resources: #UfoResources object
* @command_queue: A valid cl_command_queue
*
* Returns: Euclidean distance between @arg1 and @arg2.
*/
gfloat
ufo_op_euclidean_distance (UfoBuffer *arg1,
UfoBuffer *arg2,
UfoResources *resources,
gpointer command_queue)
{
UfoRequisition arg1_requisition, arg2_requisition;
guint length;
gfloat diff;
gfloat norm = 0;
guint length1 = 0;
guint length2 = 0;
gfloat *values1;
gfloat *values2;
ufo_buffer_get_requisition (arg1, &arg1_requisition);
ufo_buffer_get_requisition (arg2, &arg2_requisition);
for (guint i = 0; i < arg1_requisition.n_dims; ++i)
length1 += (guint)arg1_requisition.dims[i];
for (guint i = 0; i < arg2_requisition.n_dims; ++i)
length2 += (guint)arg2_requisition.dims[i];
if (length2 != length1)
g_warning ("Sizes of buffers are not the same. Zero-padding applied.");
length = length2 < length1 ? length2 : length1;
values1 = ufo_buffer_get_host_array (arg1, command_queue);
values2 = ufo_buffer_get_host_array (arg2, command_queue);
for (guint i = 0; i < length; ++i) {
diff = values1[i] - values2[i];
norm += powf (diff, 2);
}
for (guint i = length; i < length2; ++i)
norm += powf (values2[i], 2);
for (guint i = length; i < length1; ++i)
norm += powf (values1[i], 2);
norm = sqrtf(norm);
return norm;
}
static void
test_convert_8_from_data (Fixture *fixture,
gconstpointer unused)
{
gfloat *host_data;
ufo_buffer_convert_from_data (fixture->buffer, fixture->data8, UFO_BUFFER_DEPTH_8U);
host_data = ufo_buffer_get_host_array (fixture->buffer, NULL);
for (guint i = 0; i < fixture->n_data; i++)
g_assert (host_data[i] == ((gfloat) fixture->data8[i]));
}
static gboolean
ufo_reduce_task_process (UfoTask *task,
UfoBuffer **inputs,
UfoBuffer *output,
UfoRequisition *requisition)
{
UfoRequisition input_req;
ufo_buffer_get_requisition(inputs[0], &input_req);
float *src = ufo_buffer_get_host_array(inputs[0], NULL);
float *out = ufo_buffer_get_host_array(output, NULL);
for (unsigned i = 0; i < requisition->dims[0]; ++i) {
for (unsigned j = 0; j < requisition->dims[1]; ++j) {
out[i + j * requisition->dims[0]] = src[2 * i + 2 * j * input_req.dims[0]];
out[i + j * requisition->dims[0]] += src[2 * i + 1 + 2 * j * input_req.dims[0]];
out[i + j * requisition->dims[0]] += src[2 * i + j * 2 * input_req.dims[0] + input_req.dims[0]];
out[i + j * requisition->dims[0]] += src[2 * i + 1 + j * 2 * input_req.dims[0] + input_req.dims[0]];
}
}
return TRUE;
}
static gboolean
ufo_averager_task_process (UfoTask *task,
UfoBuffer **inputs,
UfoBuffer *output,
UfoRequisition *requisition)
{
UfoAveragerTaskPrivate *priv;
gfloat *in_array;
gfloat *out_array;
gsize n_pixels;
priv = UFO_AVERAGER_TASK_GET_PRIVATE (UFO_AVERAGER_TASK (task));
n_pixels = requisition->dims[0] * requisition->dims[1];
in_array = ufo_buffer_get_host_array (inputs[0], NULL);
out_array = ufo_buffer_get_host_array (output, NULL);
for (gsize i = 0; i < n_pixels; i++)
out_array[i] += in_array[i];
priv->counter++;
return TRUE;
}
static void
read_data (UfoTiffReaderPrivate *priv,
UfoBuffer *buffer,
UfoRequisition *requisition,
guint16 bits,
guint roi_y,
guint roi_height,
guint roi_step)
{
gchar *dst;
gsize step;
gsize offset;
step = requisition->dims[0] * bits / 8;
dst = (gchar *) ufo_buffer_get_host_array (buffer, NULL);
offset = 0;
if (requisition->n_dims == 3) {
/* RGB data */
gchar *src;
gsize plane_size;
plane_size = step * roi_height / roi_step;
src = g_new0 (gchar, step * 3);
for (guint i = roi_y; i < roi_y + roi_height; i += roi_step) {
guint xd = 0;
guint xs = 0;
TIFFReadScanline (priv->tiff, src, i, 0);
for (; xd < requisition->dims[0]; xd += 1, xs += 3) {
dst[offset + xd] = src[xs];
dst[offset + plane_size + xd] = src[xs + 1];
dst[offset + 2 * plane_size + xd] = src[xs + 2];
}
offset += step;
}
g_free (src);
}
else {
for (guint i = roi_y; i < roi_y + roi_height; i += roi_step) {
TIFFReadScanline (priv->tiff, dst + offset, i, 0);
offset += step;
}
}
}
static
void handle_get_result (UfoDaemon *daemon)
{
UfoDaemonPrivate *priv = UFO_DAEMON_GET_PRIVATE (daemon);
UfoBuffer *buffer;
gsize size;
buffer = ufo_output_task_get_output_buffer (UFO_OUTPUT_TASK (priv->output_task));
size = ufo_buffer_get_size (buffer);
UfoMessage *response = ufo_message_new (UFO_MESSAGE_ACK, size);
response->data = ufo_buffer_get_host_array (buffer, NULL);
response->data_size = size;
ufo_messenger_send_blocking (priv->msger, response, NULL);
ufo_output_task_release_output_buffer (UFO_OUTPUT_TASK (priv->output_task), buffer);
}
static gboolean
ufo_null_task_process (UfoTask *task,
UfoBuffer **inputs,
UfoBuffer *output,
UfoRequisition *requisition)
{
UfoNullTaskPrivate *priv;
priv = UFO_NULL_TASK_GET_PRIVATE (task);
if (priv->force_download) {
gfloat *host_array;
host_array = ufo_buffer_get_host_array (inputs[0], NULL);
host_array[0] = 0.0;
}
return TRUE;
}
/**
* ufo_op_l1_norm:
* @arg: A #UfoBuffer
* @resources: #UfoResources object
* @command_queue: A valid cl_command_queue
*
* Returns: L1 norm.
*/
gfloat
ufo_op_l1_norm (UfoBuffer *arg,
UfoResources *resources,
gpointer command_queue)
{
UfoRequisition arg_requisition;
gfloat *values;
gfloat norm = 0;
ufo_buffer_get_requisition (arg, &arg_requisition);
values = ufo_buffer_get_host_array (arg, command_queue);
for (guint i = 0; i < arg_requisition.dims[0]; ++i) {
for (guint j = 0; j < arg_requisition.dims[1]; ++j) {
norm += (gfloat) fabs (values[i * arg_requisition.dims[1] + j]);
}
}
return norm;
}
static void
ufo_edf_reader_read (UfoReader *reader,
UfoBuffer *buffer,
UfoRequisition *requisition,
guint roi_y,
guint roi_height,
guint roi_step)
{
UfoEdfReaderPrivate *priv;
gsize num_bytes;
gsize num_read;
gssize offset;
gchar *data;
priv = UFO_EDF_READER_GET_PRIVATE (reader);
data = (gchar *) ufo_buffer_get_host_array (buffer, NULL);
/* size of the image width in bytes */
const gsize width = requisition->dims[0] * priv->bytes_per_sample;
const guint num_rows = requisition->dims[1];
const gsize end_position = ftell (priv->fp) + priv->height * width;
offset = 0;
/* Go to the first desired row */
fseek (priv->fp, roi_y * width, SEEK_CUR);
if (roi_step == 1) {
/* Read the full ROI at once if no stepping is specified */
num_bytes = width * roi_height;
num_read = fread (data, 1, num_bytes, priv->fp);
if (num_read != num_bytes)
return;
}
else {
for (guint i = 0; i < num_rows - 1; i++) {
num_read = fread (data + offset, 1, width, priv->fp);
if (num_read != width)
return;
offset += width;
fseek (priv->fp, (roi_step - 1) * width, SEEK_CUR);
}
/* Read the last row without moving the file pointer so that the fseek to
* the image end works properly */
num_read = fread (data + offset, 1, width, priv->fp);
if (num_read != width)
return;
}
/* Go to the image end to be in a consistent state for the next read */
fseek (priv->fp, end_position, SEEK_SET);
if ((G_BYTE_ORDER == G_LITTLE_ENDIAN) && priv->big_endian) {
guint32 *conv = (guint32 *) ufo_buffer_get_host_array (buffer, NULL);
guint n_pixels = requisition->dims[0] * requisition->dims[1];
for (guint i = 0; i < n_pixels; i++)
conv[i] = g_ntohl (conv[i]);
}
}
static gboolean
ufo_transpose_task_process (UfoTask *task,
UfoBuffer **inputs,
UfoBuffer *output,
UfoRequisition *requisition)
{
gfloat *host_array;
gfloat *transposed;
gint width = (gint) requisition->dims[0];
gint height = (gint) requisition->dims[1];
gint fast_width = width - width % 4;
gint fast_height = height - height % 4;
const gint block_size = 128;
host_array = ufo_buffer_get_host_array (inputs[0], NULL);
transposed = ufo_buffer_get_host_array (output, NULL);
#ifdef __SSE__
/* Use SSE to do a 4x4 micro transposition, execute such transpositions in a
* block-based fashion and last, execute as many blocks as fit the image
* dimensions reduced by modulo 4 outliers. */
#pragma omp parallel for
for (gint j = 0; j < fast_height; j += block_size) {
guint block_j = j + block_size < fast_height ? j + block_size : fast_height;
for (gint i = 0; i < fast_width; i += block_size) {
guint block_i = i + block_size < fast_width ? i + block_size : fast_width;
for (guint l = j; l < block_j; l+=4) {
for (guint k = i; k < block_i; k+=4) {
transpose_sse (host_array + k * height + l, transposed + l * width + k, width, height);
}
}
}
}
/* Finish the outliers which exceed block size. This loop cannot be executed
* in parallel to the previous one, otherwise the vector access and outlier
* access might happen at the same time producing invalid results. */
#pragma omp parallel for
for (gint j = 0; j < height; j++) {
/* If we are in the height which was processed in a vectorized way treat
* only the x outlier, otherwise the whole row has to be processed. */
guint start_i = j < fast_height ? fast_width : 0;
for (gint i = start_i; i < width; i++) {
transposed[j * width + i] = host_array[i * height + j];
}
}
#else
/* Transpose in blocks */
#pragma omp parallel for
for (gint j = 0; j < height; j += block_size) {
guint block_j = j + block_size < height ? j + block_size : height;
for (gint i = 0; i < width; i += block_size) {
guint block_i = i + block_size < width ? i + block_size : width;
for (guint l = j; l < block_j; l++) {
for (guint k = i; k < block_i; k++) {
transposed[l * width + k] = host_array[k * height + l];
}
}
}
}
#endif
return TRUE;
}
static void
test_create_lots_of_buffers (Fixture *fixture,
gconstpointer unused)
{
gint num_buffers = 10000;
UfoConfig *config = ufo_config_new ();
UfoResources *resources = ufo_resources_new (config, NULL);
gpointer context = ufo_resources_get_context (resources);
UfoRequisition requisition = {
.n_dims = 2,
.dims[0] = 800,
.dims[1] = 800
};
while (num_buffers-- > 0) {
UfoBuffer *buffer = ufo_buffer_new (&requisition, NULL, context);
gpointer device_array = ufo_buffer_get_device_array (buffer, NULL);
ufo_buffer_discard_location (buffer);
g_assert (device_array != NULL);
g_object_unref (buffer);
}
}
static void
test_copy_lots_of_buffers (Fixture *fixture,
gconstpointer *unused)
{
gint num_buffers = 10000;
UfoConfig *config = ufo_config_new ();
UfoResources *resources = ufo_resources_new (config, NULL);
gpointer context = ufo_resources_get_context (resources);
UfoRequisition requisition = {
.n_dims = 2,
.dims[0] = 800,
.dims[1] = 800
};
// TODO enforce copy between GPUs
UfoBuffer *src = ufo_buffer_new (&requisition, NULL, context);
UfoBuffer *dest = ufo_buffer_new (&requisition, NULL, context);
while (num_buffers-- > 0) {
g_debug("copy");
ufo_buffer_copy (src, dest);
}
}
static void
test_convert_8 (Fixture *fixture,
gconstpointer unused)
{
gfloat *host_data;
host_data = ufo_buffer_get_host_array (fixture->buffer, NULL);
g_memmove (host_data, fixture->data8, fixture->n_data);
ufo_buffer_convert (fixture->buffer, UFO_BUFFER_DEPTH_8U);
host_data = ufo_buffer_get_host_array (fixture->buffer, NULL);
for (guint i = 0; i < fixture->n_data; i++)
g_assert (host_data[i] == ((gfloat) fixture->data8[i]));
}
