在OpenCL内核中读取GL_UNSIGNED_BYTE OpenGL texture2D(android)

我的
Android应用程序将OpenGL texture2D传递给我的OpenCL内核,但是我的内核读取的像素值超出界限(> 255).

我像这样创建我的OpenGL纹理:

    GLES20.glGenTextures ( 2, targetTex, 0 );
    GLES20.glBindTexture(GLES20.GL_TEXTURE_2D, targetTex[0]);
    GLES20.glTexParameteri(GLES20.GL_TEXTURE_2D, GLES20.GL_TEXTURE_MIN_FILTER, GLES20.GL_LINEAR);
    GLES20.glTexParameteri(GLES20.GL_TEXTURE_2D, GLES20.GL_TEXTURE_MAG_FILTER, GLES20.GL_LINEAR);
    GLES20.glTexImage2D(GLES20.GL_TEXTURE_2D, 0, GLES20.GL_RGBA, image_width, image_height, 0, GLES20.GL_RGBA, GLES20.GL_UNSIGNED_BYTE, null);
    GLES20.glBindTexture(GLES20.GL_TEXTURE_2D, 0);

然后通过将其绑定到FBO来渲染纹理:

    targetFramebuffer = IntBuffer.allocate(1);
    GLES20.glGenFramebuffers(1, targetFramebuffer);
    GLES20.glBindFramebuffer(GLES20.GL_FRAMEBUFFER, targetFramebuffer.get(0));
    GLES20.glFramebufferTexture2D(GLES20.GL_FRAMEBUFFER, GLES20.GL_COLOR_ATTACHMENT0, GLES20.GL_TEXTURE_2D, targetTex[0], 0);
    GLES20.glBindFramebuffer(GLES20.GL_FRAMEBUFFER, 0);

我像这样创建cl内存对象:

    mem_images[0] = clCreateFromGLTexture2D(m_clContext, CL_MEM_READ_ONLY, GL_TEXTURE_2D, 0, in_tex, &err);

这是我的OpenCL内核:

const sampler_t sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_NONE | CLK_FILTER_NEAREST;

kernel void transfer_data(__read_only image2d_t input_image, __global float* debug) {
    int2 pos;
    uint4 pixel;
    for (pos.y = get_global_id(1); pos.y < HEIGHT; pos.y += get_global_size(1)) {
        for (pos.x = get_global_id(0); pos.x < WIDTH; pos.x += get_global_size(0)) {
            pixel = read_imageui(input_image, sampler, pos);
            debug[(pos.x + pos.y*WIDTH)*NUM_CHANNELS + 0] = pixel.x;
            debug[(pos.x + pos.y*WIDTH)*NUM_CHANNELS + 1] = pixel.y;
            debug[(pos.x + pos.y*WIDTH)*NUM_CHANNELS + 2] = pixel.z;        
        }
    }
}

这就是我对内核进行排队的方法:

    local2Dsize[0] = 4;
    local2Dsize[1] = 4;
    global2Dsize[0] = clamp(image_width, 0, max_work_items[0]);
    global2Dsize[1] = clamp(image_height, 0, max_work_items[1]);

    global2Dsize[0] = ceil((float)global2Dsize[0]/(float)local2Dsize[0])*local2Dsize[0];
    global2Dsize[1] = ceil((float)global2Dsize[1]/(float)local2Dsize[1])*local2Dsize[1];

    twoDlocal_sizes["transfer_data"] = local2Dsize;
    twoDglobal_sizes["transfer_data"] = global2Dsize;

    kernels["transfer_data"] = clCreateKernel(m_program, "transfer_data", &err);

    err  = clSetKernelArg(kernels["transfer_data"], 0, sizeof(cl_mem), &mem_images[0]);
    err |= clSetKernelArg(kernels["transfer_data"], 1, sizeof(cl_mem), &mems["image"]);

err = clEnqueueAcquireGLObjects(m_queue, 1, &mem_images[0], 0, 0, 0);
CHECK_ERROR_OCL(err, "acquiring GL objects", return false);

    err = clEnqueueNDRangeKernel(m_queue, kernels["transfer_data"], 2, NULL, twoDglobal_sizes["transfer_data"], twoDlocal_sizes["transfer_data"], 0, NULL, NULL);

    err = clFinish(m_queue);

err = clEnqueueReleaseGLObjects(m_queue, 1, &mem_images[0], 0, 0, 0);
CHECK_ERROR_OCL(err, "releasing GL objects", return false);

现在回到主机上,当我打印出这些像素值(来自阵列调试)时,它们已经超出界限,我不明白为什么会这样.

如果您需要更多见解:

我的项目的总体目标是以OpenGL纹理的形式获取相机帧,使用OpenCL处理它们并将输出渲染回屏幕.
但是从android相机获得的纹理只能绑定到GL_TEXTURE_EXTERNAL_OES(http://developer.android.com/reference/android/graphics/SurfaceTexture.html),这不是一个有效的纹理来创建OpenCL内存对象.
因此,我将相机输出渲染到GL_TEXTURE_2D并将其传递给OpenCL.

我确信像素正在正确地渲染到纹理,因为当我在屏幕上显示纹理时(不涉及任何OpenCL),它会正确显示图像.

我通过创建纹理(而不是从相机获取数据)并将其传递给opencl来进行一些调试.所以这些是我得到的映射:

    0 -> 0
    1 -> 7172
    2 -> 8196
    3 -> 8710
    4 -> 9220
    5 -> 9477
    6 -> 9734
    7 -> 9991
    8 -> 10244
    9 -> 10372
    10 -> 10501
    11 -> 10629
    12 -> 10758
    13 -> 10886
    14 -> 11015
    15 -> 11143
    16 -> 11268
    17 -> 11332
    18 -> 11396
    19 -> 11460
    20 -> 11525
    21 -> 11589
    22 -> 11653
    23 -> 11717
    24 -> 11782
    25 -> 11846
    26 -> 11910
    27 -> 11974
    28 -> 12039
    29 -> 12103
    30 -> 12167
    31 -> 12231
    32 -> 12292
    33 -> 12324
    34 -> 12356
    35 -> 12388
    36 -> 12420
    37 -> 12452
    38 -> 12484
    39 -> 12516
    40 -> 12549
    41 -> 12581
    42 -> 12613
    43 -> 12645
    44 -> 12677
    45 -> 12709
    46 -> 12741
    47 -> 12773
    48 -> 12806
    49 -> 12838
    50 -> 12870
    51 -> 12902
    52 -> 12934
    53 -> 12966
    54 -> 12998
    55 -> 13030
    56 -> 13063
    57 -> 13095
    58 -> 13127
    59 -> 13159
    60 -> 13191
    61 -> 13223
    62 -> 13255
    63 -> 13287
    64 -> 13316
    65 -> 13332
    66 -> 13348
    67 -> 13364
    68 -> 13380
    69 -> 13396
    70 -> 13412
    71 -> 13428
    72 -> 13444
    73 -> 13460
    74 -> 13476
    75 -> 13492
    76 -> 13508
    77 -> 13524
    78 -> 13540
    79 -> 13556
    80 -> 13573
    81 -> 13589
    82 -> 13605
    83 -> 13621
    84 -> 13637
    85 -> 13653
    86 -> 13669
    87 -> 13685
    88 -> 13701
    89 -> 13717
    90 -> 13733
    91 -> 13749
    92 -> 13765
    93 -> 13781
    94 -> 13797
    95 -> 13813
    96 -> 13830
    97 -> 13846
    98 -> 13862
    99 -> 13878
    100 -> 13894
    101 -> 13910
    102 -> 13926
    103 -> 13942
    104 -> 13958
    105 -> 13974
    106 -> 13990
    107 -> 14006
    108 -> 14022
    109 -> 14038
    110 -> 14054
    111 -> 14070
    112 -> 14087
    113 -> 14103
    114 -> 14119
    115 -> 14135
    116 -> 14151
    117 -> 14167
    118 -> 14183
    119 -> 14199
    120 -> 14215
    121 -> 14231
    122 -> 14247
    123 -> 14263
    124 -> 14279
    125 -> 14295
    126 -> 14311
    127 -> 14327
    128 -> 14340
    129 -> 14348
    130 -> 14356
    131 -> 14364
    132 -> 14372
    133 -> 14380
    134 -> 14388
    135 -> 14396
    136 -> 14404
    137 -> 14412
    138 -> 14420
    139 -> 14428
    140 -> 14436
    141 -> 14444
    142 -> 14452
    143 -> 14460
    144 -> 14468
    145 -> 14476
    146 -> 14484
    147 -> 14492
    148 -> 14500
    149 -> 14508
    150 -> 14516
    151 -> 14524
    152 -> 14532
    153 -> 14540
    154 -> 14548
    155 -> 14556
    156 -> 14564
    157 -> 14572
    158 -> 14580
    159 -> 14588
    160 -> 14597
    161 -> 14605
    162 -> 14613
    163 -> 14621
    164 -> 14629
    165 -> 14637
    166 -> 14645
    167 -> 14653
    168 -> 14661
    169 -> 14669
    170 -> 14677
    171 -> 14685
    172 -> 14693
    173 -> 14701
    174 -> 14709
    175 -> 14717
    176 -> 14725
    177 -> 14733
    178 -> 14741
    179 -> 14749
    180 -> 14757
    181 -> 14765
    182 -> 14773
    183 -> 14781
    184 -> 14789
    185 -> 14797
    186 -> 14805
    187 -> 14813
    188 -> 14821
    189 -> 14829
    190 -> 14837
    191 -> 14845
    192 -> 14854
    193 -> 14862
    194 -> 14870
    195 -> 14878
    196 -> 14886
    197 -> 14894
    198 -> 14902
    199 -> 14910
    200 -> 14918
    201 -> 14926
    202 -> 14934
    203 -> 14942
    204 -> 14950
    205 -> 14958
    206 -> 14966
    207 -> 14974
    208 -> 14982
    209 -> 14990
    210 -> 14998
    211 -> 15006
    212 -> 15014
    213 -> 15022
    214 -> 15030
    215 -> 15038
    216 -> 15046
    217 -> 15054
    218 -> 15062
    219 -> 15070
    220 -> 15078
    221 -> 15086
    222 -> 15094
    223 -> 15102
    224 -> 15111
    225 -> 15119
    226 -> 15127
    227 -> 15135
    228 -> 15143
    229 -> 15151
    230 -> 15159
    231 -> 15167
    232 -> 15175
    233 -> 15183
    234 -> 15191
    235 -> 15199
    236 -> 15207
    237 -> 15215
    238 -> 15223
    239 -> 15231
    240 -> 15239
    241 -> 15247
    242 -> 15255
    243 -> 15263
    244 -> 15271
    245 -> 15279
    246 -> 15287
    247 -> 15295
    248 -> 15303
    249 -> 15311
    250 -> 15319
    251 -> 15327
    252 -> 15335
    253 -> 15343
    254 -> 15351
    255 -> 15359

左边是我在OpenGL纹理中输入的颜色值,左边是我在OpenCL中读取值时得到的相应值.

最佳答案 GL_UNSIGNED_BYTE纹理作为CL_UNORM_INT8映射到OpenCL.您需要使用read_imagef来读取这些图像而不是read_imageui.使用read_imageui时看到的值是内部浮点格式的原始位.

点赞