Archives Discussions

[zoomzoom]

I modified vector_copy.c sample from CLOC 0.9 to query the alignment for the kernel symbol segment and it returns the value 3:

   /*

    * Extract the symbol from the executable.

    */

    hsa_executable_symbol_t symbol;

    err = hsa_executable_get_symbol(executable, "", "&__OpenCL_vector_copy_kernel", agent, 0, &symbol);

    check(Extract the symbol from the executable, err);

    /*

     * Extract dispatch information from the symbol

     */

    uint32_t kernarg_segment_align;

    err = hsa_executable_symbol_get_info(symbol, HSA_EXECUTABLE_SYMBOL_INFO_KERNEL_KERNARG_SEGMENT_ALIGNMENT, &kernarg_segment_align);

    printf("\nKERNARG_SEGMENT_ALIGNMENT = %d", kernarg_segment_align);

According to the HSA Runtime Specs, this value should be: "the maximum of 16 and the maximum alignment of any of the kernel arguments."

Am I doing something wrong or is the returned value incorrect?

Thanks.

[jedwards]

The return value is incorrect. The value should be 16. I have opened a issue on this problem and it will be corrected in a future release.

[jedwards · Accepted solution]

This is fixed in release version release-v1.0f.2.