Пример #1
0
int main(){
	double temperature = 0;		// 気温
	char unit;					// 単位
	double result;

	cout << "気温を入力してください。(c: 摂氏, f: 華氏)" << endl;

	while(cin >> temperature >> unit){
		switch(unit){
			case 'c':
				if(!int_check(ctof(temperature)))
					error("int型で表すことができません。");
				cout << " -> " << (int)ctof(temperature) << "℉" << endl;
				break;
			case 'f':
				if(!int_check(ftoc(temperature)))
					error("int型で表すことができません。");
				cout << " -> " << (int)ftoc(temperature) << "℃" << endl;
				break;
			default:
				cout << "そのような単位は入力できません。" << endl;
				break;
		}		
	}
}
Пример #2
0
int main()
{
	try {
		cout << "This program can convert Celsius to Fahrenheit and vice-versa.\n";
		cout << "To convert fahrenheit to celsius type in: 45f or 45c for celsius to fahrenheit.\n";
		double temp;
		string temp_type;
		cin >> temp >> temp_type;

		if (temp_type == "c") {
			double f = ctof(temp);
			cout << "The fahrenheit equivalent of " << temp << " celsius is: " << f << endl;
		}
		else if (temp_type == "f") {
			double c = ftoc(temp);
			cout << "The celsius equivalent of " << temp << " fahrenheit is: " << c << endl;
		}
		else {
			cerr << "Invalid input.\n";
			return 1;
		}
	}
	catch(runtime_error e) {
		cerr << e.what();
	}
	return 0;
}
Пример #3
0
int main() {
  const uint8_t lower = 0;
  const uint16_t upper = 300;
  const uint8_t step = 20;

  puts("-- TEMP CONVERSION PROGRAM (C to F) --\n");
  for (double celcius = lower; celcius <= upper; celcius += step)
    printf("%3.0f %6.1f\n", celcius, ctof(celcius));

  return 0;
}
int main()
try
{
	cout << "Enter a temperature and unit of temperature (c = Celsius, f = Fahrenheit)\n";

	double temp_to_convert = 0; //input temperature variable
	cin >> temp_to_convert;
	if (cin.fail()) {
		cin.clear();  //clear cin error flags
		cin.ignore(INT_MAX, '\n'); //clear cin buffer
		error("entered non-numeric temperature!\n");
	}

	char entered_temp_unit = '?';
	cin >> entered_temp_unit;
	entered_temp_unit = tolower(entered_temp_unit);

	double converted_temp = 0;
	char converted_temp_unit = '?';
	switch (entered_temp_unit) {
	case 'f' : 
		converted_temp = ftoc(temp_to_convert);
		converted_temp_unit = 'c';
		break;
	case 'c' : 
		converted_temp = ctof(temp_to_convert);
		converted_temp_unit = 'f';
		break;
	default : 
		error("entered wrong temperature unit of measure!\n");
	}
	cout << "Converted temperature is " << converted_temp << " " << converted_temp_unit << '\n';
	
	keep_window_open();
	return 0;
}
catch (exception& e)
{
	cerr << "error: " << e.what() << '\n';
	keep_window_open();
	return 1;
}
catch (...)
{
	cerr << "Oops: unknown exception!\n";
	keep_window_open();
	return 2;
}
int main()
{
  try
    {
      double c = 0; // decplare input variable
      if(std::cin >> c)// retrieve temperature into input variable
	{
	  //double k = ctok(c); // celsius to kelvin
	  //double k = ktoc(c); // kelvin to celsius
	  double k = ctof(c);
	  std::cout << k << std::endl ; // print out temperature
	}
      else
	{
	  std::cerr << "invalid input " << std::endl;
	}
    }
Пример #6
0
int main(int argc, char* argv[])
{
    if (argc != 3) {

        print_help(argv[0]);
        exit(EXIT_FAILURE);
    }

    char* endptr = NULL;

    errno = 0;
    float temp = strtof(argv[2], &endptr);

    if (errno || *endptr)
        die("Your temperature is not a valid number.");

    puts("Your argument converts to:");
    switch (argv[1][1]) {

    case 'c':
        printf("%.2fK\n", ctok(temp));
        printf("%.2fF\n", ctof(temp));

        break;

    case 'f':
        printf("%.2fC\n", ftoc(temp));
        printf("%.2fK\n", ftok(temp));

        break;

    case 'k':
        printf("%.2fC\n", ktoc(temp));
        printf("%.2fF\n", ktof(temp));

        break;

    default:
        die("Unknown temperature type");
    }

    return EXIT_SUCCESS;
}
Пример #7
0
/*NP*/
void write_file2( char *filename )
/*************************************************************/
/*  WRITE_FILE                                               */
/*  John Hart  NSSFC KCMO                                    */
/*                                                           */
/*  Writes contents of sndg array into SHARP95 file.         */
/*************************************************************/
{
    short i, j;
    short idx[7];
    float sfctemp, sfcdwpt, sfcpres, j1, j2, ix1;
    struct _parcel pcl;
    char st[80];
    FILE *fout;

    idx[1]  = getParmIndex("PRES");
    idx[2]  = getParmIndex("HGHT");
    idx[3]  = getParmIndex("TEMP");
    idx[4]  = getParmIndex("DWPT");
    idx[5]  = getParmIndex("DRCT");
    idx[6]  = getParmIndex("SPED");

    fout = fopen( filename, "wt" );
    if (fout==NULL)
    {
        printf("Unable to write output file!\n" );
        return;
    }
    fputs( "%TITLE%\n", fout );
    fputs( raobtitle, fout );
    fputs( "\n\n", fout );
    fprintf( fout, "   LEVEL       HGHT       TEMP       DWPT       WDIR       WSPD\n");
    fprintf( fout, "-------------------------------------------------------------------\n");
    fputs( "%RAW%\n", fout );
    for(i=0; i<numlvl; i++)
    {
        for(j=1; j<=5; j++) fprintf( fout, "%8.2f,  ", sndg[i][idx[j]]);
        fprintf( fout, "%8.2f\n", sndg[i][idx[6]]);
    }
    fputs( "%END%\n\n", fout );

    if ((numlvl<4) || (!qc(i_dwpt(700, I_PRES)))) return;

    fprintf( fout, "----- Parcel Information-----\n");

    /* ----- Calculate Parcel Data ----- */
    sfctemp = lplvals.temp;
    sfcdwpt = lplvals.dwpt;
    sfcpres = lplvals.pres;

    strcpy( st, "*** " );
    strcat( st, lplvals.desc );
    strcat( st, " ***" );
    fprintf( fout, "%s\n", st);
    ix1 = parcel( -1, -1, sfcpres, sfctemp, sfcdwpt, &pcl);

    fprintf( fout, "LPL:  P=%.0f  T=%.0fF  Td=%.0fF\n", sfcpres, ctof(sfctemp), ctof(sfcdwpt));
    fprintf( fout, "CAPE:            %6.0f J/kg\n", pcl.bplus);
    fprintf( fout, "CINH:            %6.0f J/kg\n", pcl.bminus);

    fprintf( fout, "LI:              %6.0f C\n", pcl.li5);
    fprintf( fout, "LI(300mb):       %6.0f C\n", pcl.li3);

    fprintf( fout, "3km Cape:        %6.0f J/kg\n", pcl.cape3km);
    j1 = pcl.bplus;
    j2 = i_hght(pcl.elpres, I_PRES) - i_hght(pcl.lfcpres, I_PRES);
    fprintf( fout, "NCAPE:           %6.2f m/s2\n\n", j1/j2);

    fprintf( fout, "LCL:    %6.0fmb     %6.0fm\n", pcl.lclpres, agl(i_hght(pcl.lclpres, I_PRES)));
    fprintf( fout, "LFC:    %6.0fmb     %6.0fm\n", pcl.lfcpres, agl(i_hght(pcl.lfcpres, I_PRES)));
    fprintf( fout, "EL:     %6.0fmb     %6.0fm\n", pcl.elpres, agl(i_hght(pcl.elpres, I_PRES)));
    fprintf( fout, "MPL:    %6.0fmb     %6.0fm\n", pcl.mplpres, agl(i_hght(pcl.mplpres, I_PRES)));
    fprintf( fout, "All heights AGL\n\n" );

    fprintf( fout, "----- Moisture -----\n" );
    strcpy( st, qc2( precip_water( &ix1, -1, -1), " in", 2 ));
    fprintf( fout, "Precip Water:    %s\n", st);
    strcpy( st, qc2( mean_mixratio( &ix1, -1, -1 ), " g/Kg", 1 ));
    fprintf( fout, "Mean W:          %s\n\n", st);

    fprintf( fout, "----- Lapse Rates -----\n" );
    j1 = delta_t(&ix1);
    j2 = lapse_rate( &ix1, 700, 500);
    fprintf( fout, "700-500mb   %.0f C      %.1f C/km\n", j1, j2);

    j1 = vert_tot(&ix1);
    j2 = lapse_rate( &ix1, 850, 500);
    fprintf( fout, "850-500mb   %.0f C      %.1f C/km\n", j1, j2);

    fclose( fout );
}
Пример #8
0
void mgv(ftype f[], ftype u[], ftype dx, unsigned n1,unsigned n2,unsigned n3, size_t field_size, unsigned points, int use_alignment, unsigned dim_x, cl_context ctx, cl_command_queue queue, cl_kernel poisson_knl, int wg_dims , int wg_x, int wg_y, int wg_z, int z_div, int fetch_per_pt, int flops_per_pt){
  // mgv does one v-cycle for the Poisson problem on a grid with mesh size dx
  // Inputs: f is right hand side, u is current approx dx is mesh size, n1 number of sweeps
  // on downward branch, n2 number of sweeps on upwardbranch, n3 number of sweeps on
  // coarsest grid.
  // Output:  It just returns an updated version of u
  cl_ulong start_big;
  #ifdef DO_TIMING
  cl_event evt;
  cl_event *evt_ptr = &evt;
  #else
  cl_event *evt_ptr = NULL;
  #endif
  size_t i, isweep;
  item * ugrid, * head, * curr;
  int l = 0;
  ftype dxval[POINTS/2] = {0};  // this is huge and unnecessary.  Try to cut downif time!!  
  ftype h;
  unsigned nx[POINTS/2] = {0};
  // --- Allocate common gpu memory----
  cl_int status;
  cl_mem dev_buf_u = clCreateBuffer(ctx, CL_MEM_READ_WRITE, field_size * sizeof(ftype), 0, &status);
  CHECK_CL_ERROR(status, "clCreateBuffer");
  cl_mem dev_buf_f = clCreateBuffer(ctx, CL_MEM_READ_ONLY, field_size * sizeof(ftype), 0, &status);
  CHECK_CL_ERROR(status, "clCreateBuffer");
  cl_mem dev_buf_hist_u = clCreateBuffer(ctx, CL_MEM_READ_ONLY, field_size * sizeof(ftype), 0, &status);
  CHECK_CL_ERROR(status, "clCreateBuffer");
  //cl_mem read_buf = clCreateBuffer(ctx, CL_MEM_READ_ONLY, field_size * sizeof(ftype), 0, &status);
  CHECK_CL_ERROR(status, "clCreateBuffer");
  // -----------------------------------
  dxval[0] = dx;
  nx[0] = points;
  //const size_t max_size  = POINTS * POINTS * ((POINTS + 15)/16) * 16;
  // --------------- Allocatig the finest grid --------------------
  ugrid = (item *)malloc(sizeof(item));
  ugrid->uvec = malloc(field_size * sizeof(ftype));
  ugrid->fvec = malloc(field_size * sizeof(ftype));
  ugrid->rvec = malloc(field_size * sizeof(ftype));
  ugrid->dim_other = nx[0];
  ugrid->dim_x = dim_x;
  for(i = 0; i < field_size; i++){
     ugrid->uvec[i] = u[i];
     ugrid->fvec[i] = f[i];
     ugrid->rvec[i] = 0;
  }
  head = ugrid;  // head will always be the first one

  // ---------------- Set up the coarse grids ----------------------
  while((nx[l] - 1) % 2 == 0 && nx[l] > 3){
    l = l+1;
    nx[l] = (nx[l - 1] - 1) / 2 + 1;
    dxval[l] = 2 * dxval[l-1]; 
    curr = (item *)malloc(sizeof(item));
    curr->uvec = malloc(field_size * sizeof(ftype));
    curr->fvec = malloc(field_size * sizeof(ftype));
    curr->rvec = malloc(field_size * sizeof(ftype));

    curr->dim_other = nx[l];
    curr->field_start = 0;
    curr->dim_x = curr->dim_other;
    if(use_alignment)
    	curr->dim_x = ((nx[l] + 15)/16) * 16;

    // initialize vectors
    for(i = 0; i < field_size; i++){
 	curr->uvec[i] = 0;
	curr->fvec[i] = 0;
     	curr->rvec[i] = 0;
    }
    ugrid->next = curr; // curr gets attached to ugrid
    curr->prev = ugrid;
    ugrid = curr;
  }
  int nl = l; // this is the maximum number of grids that were created
  // --- at this point head contains the finest grid and ugrid contains the coarsest -----
  curr = head;
  head->prev = NULL;
  ugrid->next = NULL;

  // ---------------- Now relax each of the different grids descending--------
  for(l = 0; l < nl; l++){  // I stop right before nl (will be treated different)
     // ----------------------------------------------------------------------
     // -------------------- GPU DESCENDING V-CYCLE --------------------------
     // ----------------------------------------------------------------------
     {
     if(curr->dim_other < CUTOFF){
	for(isweep = 0; isweep < n1; isweep++){
	     gsrelax(curr, dxval[l]);
	}
     }

     else{
  	// ---GPU------GPU------GPU------GPU------GPU------GPU------GPU------GPU--- //
  	// fill in the buffers inside the GPU with the current data
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_u, CL_TRUE, 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_f, CL_TRUE, 0, field_size * sizeof(ftype), curr->fvec, 0, NULL, NULL));
	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_hist_u, CL_TRUE, 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
  	h = dxval[l] * dxval[l];
        size_t gdim[] = { curr->dim_x-16, curr->dim_x-16, curr->dim_x/z_div };
        size_t ldim[] = { wg_x, wg_y, wg_z };

  	for(i = 0; i < n1; i++){
     	   // ----------------------------------------------------------------------
     	   // invoke poisson kernel
     	   // ----------------------------------------------------------------------
	   //size_t u_size;
	   //CALL_CL_GUARDED(clGetMemObjectInfo, (dev_buf_u, CL_MEM_SIZE, sizeof(u_size), &u_size, 0));
	   //int u_size_i = u_size;
	   //printf("u_size=%d fstart=%d dim_x=%d dim_other=%d\n" , u_size_i, curr->field_start, curr->dim_x, curr->dim_other);
	   curr->field_start = 0;
     	   SET_7_KERNEL_ARGS(poisson_knl, dev_buf_u, dev_buf_f, dev_buf_hist_u, curr->field_start, curr->dim_x, curr->dim_other, h);
     	   // run the kernel
     	   CALL_CL_GUARDED(clEnqueueNDRangeKernel, (queue, poisson_knl, /*dimensions*/ wg_dims, NULL, gdim, ldim, 0, NULL, evt_ptr));
     	   #ifdef DO_TIMING
      	   // If timing is enabled, this wait can mean a significant performance hit.
      	   CALL_CL_GUARDED(clWaitForEvents, (1, &evt));
 
      	   cl_ulong start, end;
      	   CALL_CL_GUARDED(clGetEventProfilingInfo, (evt, CL_PROFILING_COMMAND_START, sizeof(start), &start, NULL));
      	   CALL_CL_GUARDED(clGetEventProfilingInfo, (evt, CL_PROFILING_COMMAND_END, sizeof(start), &end, NULL));
 
      	   gbytes_accessed += 1e-9*(sizeof(ftype) * field_size * fetch_per_pt);
	   start_big = start;
      	   seconds_taken += 1e-9*(end-start);
      	   mcells_updated += curr->dim_other*curr->dim_other*curr->dim_other/1e6;
      	   gflops_performed += 1e-9*curr->dim_x*curr->dim_x*curr->dim_x * flops_per_pt;

      	   CALL_CL_GUARDED(clReleaseEvent, (evt));
     	   #endif
      	   CALL_CL_GUARDED(clFinish, (queue)); //ira adentro??
	   cl_mem tmp = dev_buf_u;
	   dev_buf_u = dev_buf_hist_u;
	   dev_buf_hist_u = tmp;
     	}
        //when I'm done, read from buffer
        CALL_CL_GUARDED(clEnqueueReadBuffer, (queue, dev_buf_u, /*blocking*/ CL_TRUE, /*offset*/ 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
      }
     }
     resid2(curr, dxval[l]);
     injf2c(curr, curr->next); //this function updates f_{i+1}
     curr = curr->next;
  }
  // ----------------------------------------------------------------------
  // --------------- GPU ON THE COARSEST GRID -----------------------------
  // ----------------------------------------------------------------------
  {
    if(curr->dim_other < CUTOFF){
	for(i = 0; i < n3; i++){
	     gsrelax(curr, dxval[nl]);
	}
    }

    else{
  	// ---GPU------GPU------GPU------GPU------GPU------GPU------GPU------GPU--- //
  	// fill in the buffers inside the GPU with the current data
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_u, CL_TRUE, 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_f, CL_TRUE, 0, field_size * sizeof(ftype), curr->fvec, 0, NULL, NULL));
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_hist_u, CL_TRUE, 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
  	h = dxval[nl] * dxval[nl];
        size_t gdim[] = { curr->dim_x - 16, curr->dim_x - 16, curr->dim_x/z_div };
        size_t ldim[] = { wg_x, wg_y, wg_z };

  	for(i = 0; i < n3; i++){
     	   // ----------------------------------------------------------------------
     	   // invoke poisson kernel
     	   // ----------------------------------------------------------------------
	   curr->field_start = 0;
     	   SET_7_KERNEL_ARGS(poisson_knl, dev_buf_u, dev_buf_f, dev_buf_hist_u,curr->field_start, curr->dim_x, curr->dim_other, h);
     	   // run the kernel
	   curr->field_start = 0;
     	   CALL_CL_GUARDED(clEnqueueNDRangeKernel, (queue, poisson_knl, /*dimensions*/ wg_dims, NULL, gdim, ldim, 0, NULL, evt_ptr));
     	   #ifdef DO_TIMING
      	   // If timing is enabled, this wait can mean a significant performance hit.
      	   CALL_CL_GUARDED(clWaitForEvents, (1, &evt));
 
      	   cl_ulong start, end;
      	   CALL_CL_GUARDED(clGetEventProfilingInfo, (evt, CL_PROFILING_COMMAND_START, sizeof(start), &start, NULL));
      	   CALL_CL_GUARDED(clGetEventProfilingInfo, (evt, CL_PROFILING_COMMAND_END, sizeof(start), &end, NULL));
 
      	   gbytes_accessed += 1e-9*(sizeof(ftype) * field_size * fetch_per_pt);
      	   seconds_taken += 1e-9*(end-start);
      	   mcells_updated += curr->dim_other*curr->dim_other*curr->dim_other/1e6;
      	   gflops_performed += 1e-9*curr->dim_x*curr->dim_x*curr->dim_x * flops_per_pt;

      	   CALL_CL_GUARDED(clReleaseEvent, (evt));
     	   #endif
      	   CALL_CL_GUARDED(clFinish, (queue)); //ira adentro??
	   cl_mem tmp = dev_buf_u;
	   dev_buf_u = dev_buf_hist_u;
	   dev_buf_hist_u = tmp;
     	}
        //when I'm done, read from buffer
        CALL_CL_GUARDED(clEnqueueReadBuffer, (queue, dev_buf_u, /*blocking*/ CL_TRUE, /*offset*/ 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
     }
  }
  // ----------------------------------------------------------------------
  // -----------Upward branch of the V-cycle ------------------------------
  // ----------------------------------------------------------------------
  for(l = nl-1; l >= 0; --l){
     ctof(curr->prev, curr, field_size); //curr->prev is the finer of the two
     free(curr->uvec);  //curr won't be needed anymore
     free(curr->fvec);
     free(curr->rvec);
     curr = curr->prev;
     curr->next = NULL;
     for(isweep = 0; isweep < n2; isweep++){
	   gsrelax(curr, dxval[l]);
     }
     // Update the grids n1 times using the GPU when necessary
     {
     if(curr->dim_other < CUTOFF){
	for(isweep = 0; isweep < n2; isweep++){
	     gsrelax(curr, dxval[l]);
	}
     }

     else{
  	// ---GPU------GPU------GPU------GPU------GPU------GPU------GPU------GPU--- //
  	// fill in the buffers inside the GPU with the current data
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_u, CL_TRUE, 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_f, CL_TRUE, 0, field_size * sizeof(ftype), curr->fvec, 0, NULL, NULL));
  	CALL_CL_GUARDED(clEnqueueWriteBuffer, (queue, dev_buf_hist_u, CL_TRUE, 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
  	h = dxval[l] * dxval[l];
        size_t gdim[] = { curr->dim_x-16, curr->dim_x-16, curr->dim_x/z_div };
        size_t ldim[] = { wg_x, wg_y, wg_z };

  	for(i = 0; i < n1; i++){
     	   // ----------------------------------------------------------------------
     	   // invoke poisson kernel
     	   // ----------------------------------------------------------------------
	   curr->field_start = 0;
     	   SET_7_KERNEL_ARGS(poisson_knl, dev_buf_u, dev_buf_f, dev_buf_hist_u, curr->field_start, curr->dim_x, curr->dim_other, h);
     	   // run the kernel
     	   CALL_CL_GUARDED(clEnqueueNDRangeKernel, (queue, poisson_knl, /*dimensions*/ wg_dims, NULL, gdim, ldim, 0, NULL, evt_ptr));
     	   #ifdef DO_TIMING
      	   // If timing is enabled, this wait can mean a significant performance hit.
      	   CALL_CL_GUARDED(clWaitForEvents, (1, &evt));
 
      	   cl_ulong start, end;
      	   CALL_CL_GUARDED(clGetEventProfilingInfo, (evt, CL_PROFILING_COMMAND_START, sizeof(start), &start, NULL));
      	   CALL_CL_GUARDED(clGetEventProfilingInfo, (evt, CL_PROFILING_COMMAND_END, sizeof(start), &end, NULL));
 
      	   gbytes_accessed += 1e-9*(sizeof(ftype) * field_size * fetch_per_pt);
      	   seconds_taken += 1e-9*(end-start);
	   tot_secs += 1e-9*(end-start_big);
      	   mcells_updated += curr->dim_other*curr->dim_other*curr->dim_other/1e6;
      	   gflops_performed += 1e-9*curr->dim_x*curr->dim_x*curr->dim_x * flops_per_pt;

      	   CALL_CL_GUARDED(clReleaseEvent, (evt));
     	   #endif
      	   CALL_CL_GUARDED(clFinish, (queue)); //ira adentro??
	   cl_mem tmp = dev_buf_u;
	   dev_buf_u = dev_buf_hist_u;
	   dev_buf_hist_u = tmp;
     	}
        //when I'm done, read from buffer
        CALL_CL_GUARDED(clEnqueueReadBuffer, (queue, dev_buf_u, /*blocking*/ CL_TRUE, /*offset*/ 0, field_size * sizeof(ftype), curr->uvec, 0, NULL, NULL));
      }
     }
  }
  // ---------- and the solution is right there in the last curr curr->uvec
  for(i = 0; i < field_size; i++)
     u[i] = curr->uvec[i];
  free(curr->uvec);
  //free(curr->fvec);
  free(curr->rvec);
  //free(ugrid->uvec);
  //free(ugrid->fvec);
  //free(ugrid->rvec);
  free(curr);
  CALL_CL_GUARDED(clReleaseMemObject, (dev_buf_u));
  CALL_CL_GUARDED(clReleaseMemObject, (dev_buf_f));
  CALL_CL_GUARDED(clReleaseMemObject, (dev_buf_hist_u));
}