/* Print code to "p" for copying "array" back from the device to the host
* in its entirety. The bounds on the extent of "array" have
* been precomputed in extract_array_info and are used in
* gpu_array_info_print_size.
*/
static __isl_give isl_printer *copy_array_from_device_global_memory(
__isl_take isl_printer *p, struct gpu_array_info *array)
{
p = isl_printer_start_line(p);
p = isl_printer_print_str(p, "cudaCheckReturn(cudaMemcpy(");
if (gpu_array_is_scalar(array))
p = isl_printer_print_str(p, "&");
p = isl_printer_print_str(p, array->name);
p = isl_printer_print_str(p, ", dev_");
p = isl_printer_print_str(p, array->name);
p = isl_printer_print_str(p, ", ");
p = gpu_array_info_print_size(p, array);
p = isl_printer_print_str(p, ", cudaMemcpyDeviceToHost));");
p = isl_printer_end_line(p);
return p;
}
/* Print code to "p" for copying "array" from the host to the device
* in its entirety. The bounds on the extent of "array" have
* been precomputed in extract_array_info and are used in
* gpu_array_info_print_size.
*/
static __isl_give isl_printer *copy_array_to_device(__isl_take isl_printer *p,
struct gpu_array_info *array)
{
if(print_device_arrays_or_not(array))
{
p = isl_printer_start_line(p);
p = isl_printer_print_str(p, "cudaCheckReturn(cudaMemcpy(dev_");
p = isl_printer_print_str(p, array->name);
p = isl_printer_print_str(p, ", ");
if (gpu_array_is_scalar(array))
p = isl_printer_print_str(p, "&");
p = isl_printer_print_str(p, array->name);
p = isl_printer_print_str(p, ", ");
p = gpu_array_info_print_size(p, array);
p = isl_printer_print_str(p, ", cudaMemcpyHostToDevice));");
p = isl_printer_end_line(p);
}
return p;
}
/* Copy "array" from the host to the device (to_host = 0) or
* back from the device to the host (to_host = 1).
*/
static __isl_give isl_printer *copy_array(__isl_take isl_printer *p,
struct gpu_array_info *array, int to_host)
{
p = isl_printer_start_line(p);
p = isl_printer_print_str(p, "openclCheckReturn(");
if (to_host)
p = isl_printer_print_str(p, "clEnqueueReadBuffer");
else
p = isl_printer_print_str(p, "clEnqueueWriteBuffer");
p = isl_printer_print_str(p, "(queue, dev_");
p = isl_printer_print_str(p, array->name);
p = isl_printer_print_str(p, ", CL_TRUE, 0, ");
p = gpu_array_info_print_size(p, array);
if (gpu_array_is_scalar(array))
p = isl_printer_print_str(p, ", &");
else
p = isl_printer_print_str(p, ", ");
p = isl_printer_print_str(p, array->name);
p = isl_printer_print_str(p, ", 0, NULL, NULL));");
p = isl_printer_end_line(p);
return p;
}
