-
Notifications
You must be signed in to change notification settings - Fork 731
New issue
Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.
By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.
Already on GitHub? Sign in to your account
[SYCL] Treat "ar" image format as PI_DEVICE_BINARY_TYPE_NATIVE #12587
Changes from 2 commits
aedcf7a
2e2dc77
f017f4a
a445c43
286c3bb
64e0fd8
f917715
908427f
08f0e39
File filter
Filter by extension
Conversations
Jump to
Diff view
Diff view
There are no files selected for viewing
Original file line number | Diff line number | Diff line change |
---|---|---|
|
@@ -685,45 +685,49 @@ static uint16_t getELFHeaderType(const unsigned char *ImgData, size_t ImgSize) { | |
sycl::detail::pi::PiDeviceBinaryType | ||
getBinaryImageFormat(const unsigned char *ImgData, size_t ImgSize) { | ||
// Top-level magic numbers for the recognized binary image formats. | ||
struct { | ||
sycl::detail::pi::PiDeviceBinaryType Fmt; | ||
const uint32_t Magic; | ||
} Fmts[] = {{PI_DEVICE_BINARY_TYPE_SPIRV, 0x07230203}, | ||
{PI_DEVICE_BINARY_TYPE_LLVMIR_BITCODE, 0xDEC04342}, | ||
// 'I', 'N', 'T', 'C' ; Intel native | ||
{PI_DEVICE_BINARY_TYPE_NATIVE, 0x43544E49}}; | ||
|
||
if (ImgSize >= sizeof(Fmts[0].Magic)) { | ||
std::remove_const_t<decltype(Fmts[0].Magic)> Hdr = 0; | ||
std::copy(ImgData, ImgData + sizeof(Hdr), reinterpret_cast<char *>(&Hdr)); | ||
|
||
// Check headers for direct formats. | ||
for (const auto &Fmt : Fmts) { | ||
if (Hdr == Fmt.Magic) | ||
return Fmt.Fmt; | ||
} | ||
auto MatchMagicNumber = [&](auto Number) { | ||
if (ImgSize < sizeof(Number)) | ||
return false; | ||
return std::memcmp(ImgData, &Number, sizeof(Number)) == 0; | ||
}; | ||
|
||
if (MatchMagicNumber(uint32_t{0x07230203})) | ||
return PI_DEVICE_BINARY_TYPE_SPIRV; | ||
|
||
if (MatchMagicNumber(uint32_t{0xDEC04342})) | ||
return PI_DEVICE_BINARY_TYPE_LLVMIR_BITCODE; | ||
|
||
if (MatchMagicNumber(uint32_t{0x43544E49})) | ||
// 'I', 'N', 'T', 'C' ; Intel native | ||
return PI_DEVICE_BINARY_TYPE_LLVMIR_BITCODE; | ||
|
||
if (MatchMagicNumber(std::array{'!', '<', 'a', 'r', 'c', 'h', '>', '\n'})) | ||
// "ar" format is used to pack binaries for multiple devices, e.g. via | ||
// | ||
// -Xsycl-target-backend=spir64_gen "-device acm-g10,acm-g11" | ||
// | ||
// option. | ||
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. Small nit; I would prefer to have the comment outside the loop as having it inside the loop without brackets makes it hard to read... Unless you're used to Python code. 😉 There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. I'll keep as-is as it matches surrounding code better. |
||
return PI_DEVICE_BINARY_TYPE_NATIVE; | ||
|
||
// Check for ELF format, size requirements include data we'll read in case of | ||
// succesful match. | ||
if (ImgSize < 18 || !MatchMagicNumber(uint32_t{0x464c457F})) | ||
return PI_DEVICE_BINARY_TYPE_NONE; | ||
There was a problem hiding this comment. Choose a reason for hiding this commentThe reason will be displayed to describe this comment to others. Learn more. Nit: Though I am generally against nested nesting, a case like this I would rather we do the ELF header checks inside the if. It means when someone want to add more binary type detection, they don't have to realize that at this point we assume the remaining binary types must be ELF. |
||
|
||
uint16_t ELFHdrType = getELFHeaderType(ImgData, ImgSize); | ||
if (ELFHdrType == 0xFF04) | ||
// OpenCL executable. | ||
return PI_DEVICE_BINARY_TYPE_NATIVE; | ||
|
||
if (ELFHdrType == 0xFF12) | ||
// ZEBIN executable. | ||
return PI_DEVICE_BINARY_TYPE_NATIVE; | ||
|
||
// Newer ZEBIN format does not have a special header type, but can instead | ||
// be identified by having a required .ze_info section. | ||
if (checkELFSectionPresent(".ze_info", ImgData, ImgSize)) | ||
return PI_DEVICE_BINARY_TYPE_NATIVE; | ||
|
||
// ELF e_type for recognized binary image formats. | ||
struct { | ||
sycl::detail::pi::PiDeviceBinaryType Fmt; | ||
const uint16_t Magic; | ||
} ELFFmts[] = {{PI_DEVICE_BINARY_TYPE_NATIVE, 0xFF04}, // OpenCL executable | ||
{PI_DEVICE_BINARY_TYPE_NATIVE, 0xFF12}}; // ZEBIN executable | ||
|
||
// ELF files need to be parsed separately. The header type ends after 18 | ||
// bytes. | ||
if (Hdr == 0x464c457F && ImgSize >= 18) { | ||
uint16_t HdrType = getELFHeaderType(ImgData, ImgSize); | ||
for (const auto &ELFFmt : ELFFmts) { | ||
if (HdrType == ELFFmt.Magic) | ||
return ELFFmt.Fmt; | ||
} | ||
// Newer ZEBIN format does not have a special header type, but can instead | ||
// be identified by having a required .ze_info section. | ||
if (checkELFSectionPresent(".ze_info", ImgData, ImgSize)) | ||
return PI_DEVICE_BINARY_TYPE_NATIVE; | ||
} | ||
} | ||
return PI_DEVICE_BINARY_TYPE_NONE; | ||
} | ||
|
||
|
There was a problem hiding this comment.
Choose a reason for hiding this comment
The reason will be displayed to describe this comment to others. Learn more.
Nit, but I am not a huge fan of manual short-circuiting.