Modified a pointer in device

Hi,
I’m a newby in CUDA and I search how to modified a pointer in a device call

#include <opencv2/highgui.hpp>
#include <opencv2/core.hpp>
#include <opencv2/imgcodecs.hpp>
#include <iostream>
#include <time.h>
#include <vector>
#include <cuda.h>
#include <cuda_runtime.h>
#include <device_functions.h>
#include <device_launch_parameters.h>
#include <opencv2/cudaarithm.hpp>
#include <opencv2/core/cuda.hpp>
#include <algorithm>
#include <functional>
#include <array>
#include <thrust/sort.h>
#include <thrust/device_vector.h>
#include <thrust/device_ptr.h>

using namespace std;
using namespace cv;

__device__ 
uchar* makeVector2(uchar *mat, uchar *vec, int col, int x, int y) {
	int ind = 0;
	printf("%p \n", &mat);
	for (int i = 0; i < 9; i++) {
		mat[i] = ind;
		ind++;
	}
	/*
	for (int i = x; i < x + 3; i++) {
		for (int j = y; j < y + 3; j++) {
			vec[ind] = mat[col * i + j];
			ind++;
		}
	}
	*/
	return (mat);
}

__global__
void addF(int n, uchar *x, float *y, uchar *mat, uchar *vec)
{
	printf("%p \n", &mat);
	mat = makeVector2(mat, vec, 3, 0, 0);
}
 
int kernel(void)
{
  int N = 10;
  uchar *x, *mat, *vec;
  float *y;
 

  cudaMallocManaged(&x, N * sizeof(uchar));
  cudaMallocManaged(&y, N * sizeof(float));
  cudaMallocManaged(&vec, 9 * sizeof(uchar));
  cudaMallocManaged(&mat, 9 * sizeof(uchar));

  Mat aux = Mat::zeros(3, 3, CV_8UC1);
  
  int ind = N;
  // initialize x and y arrays on the host
  for (int i = 0; i < 3; i++) {
	  for (int j = 0; j < 3; j++) {
		  aux.at<unsigned char>(i, j) = ind;
		  ind--;
	  }
  }
  
  mat = aux.data;

  cout << &mat << endl;

  addF <<<1, 1>>>(N, x, y, mat, vec);

  cudaDeviceSynchronize();

  cout << "cactus" << endl;
  cout << &mat << endl;

  //cout << vec[0] << endl;
  
  for (int i = 0; i < 9; i++) {
	  cout << (int) mat[i] << endl;
  }
  

  // Free memory
  cudaFree(x);
  cudaFree(y);
  cudaFree(mat);
  cudaFree(vec);

  return 0;
}

When I try this mat isn’t modified and I don’t understand why.
Furthermore I use managed memory so why the pointer address isn’t the same the host and device.

Thanks

Anybody ?

You overwrote the unified memory pointer in line 72

If you removed the OpenCV dependency it might make it easier for others to help you. I’m not going to install and wrestle with OpenCV just to look at a CUDA question that doesn’t have any dependency on OpenCV.

A few observations.

  1. mat is a stack variable:
uchar *x, *mat, *vec;

You are printing out the address of mat. I wouldn’t expect that to ever change, for the life of your program:

cout << &mat << endl;
  1. your handling of mat appears broken to me. Here, you assign a value to mat:
cudaMallocManaged(&mat, 9 * sizeof(uchar));

Here, you overwrite that value, as best as I can tell:

mat = aux.data;

(your managed pointer is lost at this point). mat now has an ordinary host pointer value, one created by your Mat aux allocation.

Then you pass that ordinary host pointer to a cuda kernel call:

addF <<<1, 1>>>(N, x, y, mat, vec);

That will almost certainly be broken if you attempt to access that pointer in the kernel code.

IMO, whenever you are asking for help, you should always use proper CUDA error checking, and run your code with cuda-memcheck, before asking others for help. Your kernel would definitely be indicating errors, which you seem to be ignoring. Even if you don’t understand the error output, it will be useful for others trying to help you.

Back to the problem at hand, you seem to have a lack of understanding of pointers in C in general. Before tackling CUDA programming, it’s a good idea to have general proficiency with C and C++ programming. Basic CUDA programming depends heavily on understanding how to use pointers.

Your code may possibly be fixed (again, I don’t have OpenCV setup, and I don’t intend to do it just for this case, so can’t run your code with or without modifications) if you simply replace this very broken step:

mat = aux.data;

with something like this:

memcpy(mat, aux.data, 9*sizeof(uchar));