线程唯一标识id问题

[Eurususu]

在使用线程id的时候，如果采用核函数注释内的形式进行计算的话，对于block_size为正方形的是正常的，但是非正方形的就有问题。不过采用现在的形式来计算id的话则不会有这样的问题，是否说这个线程id的计算不能根据坐标来计算，而是要根据tid bid的形式来计算得到？

#include <iostream>
#include <opencv2/opencv.hpp>

void Error_check(cudaError_t error, const char* filename, int lineNum){
    if (error!=cudaSuccess){
        printf("cuda error:\r\ncode=%d, name=%s, description=%s\r\nfile=%s, line%d\r\n",
                error, cudaGetErrorName(error), cudaGetErrorString(error), filename, lineNum);
    }
    else{
        printf("no cuda_errors\n");
    }
}

__global__ void bgrtogray(u_char* input, u_char* output, int width, int height){
    // const int x = threadIdx.x + blockIdx.x * blockDim.x;
    // const int y = threadIdx.y + blockIdx.y * blockDim.x;
    // const int idx = x + y * width;

    const int tid = threadIdx.x + threadIdx.y * blockDim.x;
    const int bid = blockIdx.x + blockIdx.y * gridDim.x;
    const int idx = tid + blockDim.x * blockDim.y * bid;
    if (idx < width*height)
    output[idx] = 0.299f * input[3 * idx] + 0.587f * input[3 * idx + 1] + 0.114f * input[3 * idx + 2];
}


int main(){
    cv::Mat image = cv::imread("/home/jia/PycharmProjects/img_process/images/3v3蓝色黄色/train/images/chn0_20230808T100503.mp4_360.jpg");
    if (image.empty()){
        std::cerr << "failed to read image" << std::endl;
        return -1;
    }
    std::cout << image.size() << std::endl;
    int height = image.rows;
    int width = image.cols;
    int channels = image.channels();
    size_t nBytes = height * width * channels * sizeof(u_char);
    u_char* image_cuda;
    u_char* output_cuda;
    Error_check(cudaMalloc((void**)&image_cuda, nBytes), __FILE__, __LINE__);
    Error_check(cudaMalloc((void**)&output_cuda, height * width * sizeof(u_char)), __FILE__, __LINE__);
    cudaMemcpy(image_cuda, image.data, nBytes, cudaMemcpyHostToDevice);
    cudaDeviceSynchronize();
    dim3 block_size(4, 1);
    dim3 grid_size((width + block_size.x -1) / block_size.x, (height + block_size.y - 1)/block_size.y);
    bgrtogray<<<grid_size, block_size>>>(image_cuda, output_cuda, width, height);
    cudaDeviceSynchronize();
    Error_check(cudaGetLastError(), __FILE__, __LINE__);
    cv::Mat output(image.size(), CV_8UC1);
    cudaMemcpy(output.data, output_cuda, height * width * sizeof(u_char), cudaMemcpyDeviceToHost);
    cudaDeviceSynchronize();
    cudaFree(image_cuda);
    cudaFree(output_cuda);
    cv::imshow("out",output);
    cv::waitKey(5000);
    return 0;
}