pciutils/signal.c

#include <stdio.h>
#include <stdlib.h>
#include <stdbool.h>
#include <string.h>
#include <pcilib.h>
#include <pcilib/bar.h>
#include <CL/cl.h>
#include <CL/cl_ext.h>
#include "ocl.h"

/* this should actually come from the distributed pcitool sources */
#include "pciDriver.h"

typedef struct {
    /* pcilib */
    pcilib_t *pci;
    uint8_t *bar;
    cl_ulong bar_phys;

    /* OpenCL */
    OclPlatform *ocl;
    cl_device_id device;
    cl_command_queue queue;
    cl_context context;
    cl_program program;
    cl_kernel kernel;
    cl_mem fpga_buffer;
    cl_mem check_buffer;
} App;


#define WR32(addr, value) *(uint32_t *) (app->bar + (addr)) = (value);
#define RD32(addr) (*(uint32_t *) (app->bar + (addr)))


/* declaration should actually come from a distributed header file */
const pcilib_board_info_t *pcilib_get_board_info (pcilib_t *);


static bool
init_pcilib (App *app)
{
    static const char *DEVICE = "/dev/fpga0";
    const pcilib_board_info_t *board;

    app->pci = pcilib_open (DEVICE, "pci");

    if (app->pci == NULL) {
        printf ("Could not open `%s'", DEVICE);
        return false;
    }

    app->bar = pcilib_map_bar (app->pci, PCILIB_BAR0);

    if (app->bar == NULL) {
        printf ("Unable to map BAR\n");
        pcilib_close (app->pci);
        return false;
    }

    board = pcilib_get_board_info (app->pci);
    app->bar_phys = board->bar_start[PCILIB_BAR0];

    return true;
}

static void
close_pcilib (App *app)
{
    pcilib_unmap_bar (app->pci, PCILIB_BAR0, (void *) app->bar);
    pcilib_close (app->pci);
}

static cl_mem
create_fpga_buffer (App *app, size_t size, cl_int *error)
{
    cl_mem buffer;
    cl_mem_flags flags;
    cl_bus_address_amd addr;

    flags = CL_MEM_EXTERNAL_PHYSICAL_AMD;
    addr.surface_bus_address = (cl_ulong) app->bar_phys;
    addr.marker_bus_address = (cl_ulong) app->bar_phys;

    return clCreateBuffer (app->context, flags, size, &addr, error);
}

static bool
init_opencl (App *app)
{
    cl_int error;
    cl_platform_id platform;

    app->ocl = ocl_new_with_queues (0, CL_DEVICE_TYPE_GPU, 0);
    app->device = ocl_get_devices (app->ocl)[0];
    app->queue = ocl_get_cmd_queues (app->ocl)[0];
    app->context = ocl_get_context (app->ocl);

    app->program = ocl_create_program_from_file (app->ocl, "kernel.cl", NULL, &error);
    OCL_CHECK_ERROR (error);

    app->kernel = clCreateKernel (app->program, "write_to_fpga", &error);
    OCL_CHECK_ERROR (error);

    app->check_buffer = clCreateBuffer (app->context, CL_MEM_WRITE_ONLY, 8, NULL, &error);
    OCL_CHECK_ERROR (error);

    app->fpga_buffer = create_fpga_buffer (app, 1024 * 64, &error);
    OCL_CHECK_ERROR (error);

    return error != CL_SUCCESS ? false : true;
}

static void
close_opencl (App *app)
{
    OCL_CHECK_ERROR (clReleaseKernel (app->kernel));
    OCL_CHECK_ERROR (clReleaseProgram (app->program));
    OCL_CHECK_ERROR (clReleaseMemObject (app->fpga_buffer));
    OCL_CHECK_ERROR (clReleaseMemObject (app->check_buffer));
    ocl_free (app->ocl);
}

static void
check_value (App *app, uint32_t addr, uint32_t expected)
{
    uint32_t value;

    value = RD32 (addr);

    if (value != expected)
        printf ("failed [%u != %u]\n", value, expected);
    else
        printf ("success\n");
}

static void
wait_on_and_release_event (cl_event event)
{
    OCL_CHECK_ERROR (clWaitForEvents (1, &event));
    OCL_CHECK_ERROR (clReleaseEvent (event));
}

static void
launch_signal (App *app)
{
    cl_event event;
    uint32_t value;
    uintptr_t addr;
    uint32_t check[2];
    size_t global_work_size;

    addr = 0x9168;

    /* try to override defaultvalue */
    value = 0xc001;

    WR32 (addr, value);
    printf ("CPU WRITE ... ");
    check_value (app, addr, value);

    value = 0xdeadf00d;
    OCL_CHECK_ERROR (clSetKernelArg (app->kernel, 0, sizeof (cl_mem), &app->fpga_buffer));
    OCL_CHECK_ERROR (clSetKernelArg (app->kernel, 1, sizeof (cl_mem), &app->check_buffer));
    OCL_CHECK_ERROR (clSetKernelArg (app->kernel, 2, sizeof (uint32_t), &addr));
    OCL_CHECK_ERROR (clSetKernelArg (app->kernel, 3, sizeof (uint32_t), &value));

    global_work_size = 1;

    OCL_CHECK_ERROR (clEnqueueNDRangeKernel (app->queue, app->kernel, 1,
                                             NULL, &global_work_size, NULL,
                                             0, NULL, &event));
    wait_on_and_release_event (event);

    /* let's see if the GPU wrote anything */
    printf ("GPU WRITE ... ");
    check_value (app, addr, value);

    /* let's see if the kernel did at least something */
    printf ("SANITY ...... ");

    check[0] = check[1] = 0;
    OCL_CHECK_ERROR (clEnqueueReadBuffer (app->queue, app->check_buffer, CL_TRUE, 0, 8, check, 0, NULL, NULL));

    if (check[0] == addr && check[1] == value)
        printf ("success\n");
    else
        printf ("failed [0x%x != 0x%x || 0x%x != 0x%x]\n", check[0], addr, check[1], value);
}

int
main (int argc, char const* argv[])
{
    App app;

    if (!init_pcilib (&app))
        return 1;

    if (!init_opencl (&app))
        return 1;

    launch_signal (&app);

    close_opencl (&app);
    close_pcilib (&app);
    return 0;
}